asynchronous - 仅为特定内核调用 cudaDeviceSynchronize()

Question

我调用内核KerA和KerB异步。内核KerC依赖于KerB整理但独立于KerA. 那么如何以cudaDeviceSynchronize()这种方式调用，这意味着KerC等待KerB完成而不是完成KerA？

Time -------------------------->
| KerA ------------------------>
| KerB ------> | KerC --------->

Answer 1

您可以使用 CUDA 流实现此目的。

如果您不使用任何流，则使用默认流（又名流“0”），并且您不会获得并发（就像cudaDeviceSynchronize()在每个 CUDA 操作之前和之后插入一样，参见这些幻灯片）。

但是，如果KerA在流 0 中运行，KerB并KerC在流 1 中运行，您将得到您想要的，即 and 之间的同步调用KerB，KerC它们是异步的 wrt KerA。您可以使用它cudaStreamSynchronize(streamid)来同步特定的流。

Time ------------------------------------>
| Stream 0: KerA ------------------------>
| Stream 1: KerB ------> | KerC --------->

我链接的幻灯片中提供了示例。您还可以查看 SDK 的simpleStreams或concurrentKernels示例。