同步两个 CUDA 流

Question

我正在使用 CUDA 流来启用异步数据传输并隐藏内存复制延迟。我有 2 个 CPU 线程和 2 个 CUDA 流：一个是“数据”流，它本质上是由第一个 CPU 线程发起的一系列 cudaMemcpyAsync 调用，另一个是执行计算内核的“计算”流。数据流正在为计算流准备批处理，因此计算流必须确保流将要处理的批处理完全加载到内存中。

我应该使用 CUDA 事件来实现这种同步还是其他一些机制？

更新：让我澄清一下为什么我不能在每个流中使用带有数据副本/计算的单独流。问题是必须按顺序处理批次，也就是说，我不能并行执行它们（当然，这可能与多个流一起执行）。但是，在处理每批时，我可以为下一批预加载数据，从而隐藏数据传输。 以罗伯特为例：

cudaMemcpyAsync( <data for batch1>, dataStream);
cudaMemcpyAsync( <data for batch2>, dataStream);
kernelForBatch1<<<..., opsStream>>>(...);
kernelForBatch2<<<..., opsStream>>>(...);

Answer 1

您当然可以使用 cuda 事件来同步流，例如使用 cudaStreamWaitEvent API 函数。然而，将所有数据副本放在一个流中并将所有内核调用放在另一个流中的想法可能不是对流的明智使用。

保证在单个流中发出的 cuda 函数（API 调用、内核调用）按顺序执行，该流中的任何 cuda 函数在该流中所有先前的 cuda 活动完成之前不会开始（即使您正在使用cudaMemcpyAsync 等调用...)

因此，流已经为您提供了一种机制，以确保在为它复制必要的数据之前不会开始内核调用。只需在数据复制之后将该内核调用放在同一个流中即可。

这样的事情应该可以处理您的同步：

cudaMemcpyAsync( <data for kernel1>, stream1);
cudaMemcpyAsync( <data for kernel2>, stream2);
kernel1<<<..., stream1>>>(...);
kernel2<<<..., stream2>>>(...);
cudaMemcpyAsync( <data from kernel1>, stream1);
cudaMemcpyAsync( <data from kernel2>, stream2);

上述所有调用都是异步的，因此假设您已经满足异步执行的其他要求（例如使用固定内存），上述所有调用都应该“排队”并立即返回。但是，kernel1 保证不会在前面的 cudaMemcpyAsync 发布到 stream1 完成之前开始，kernel2 和 stream2 中的数据传输也是如此。

我也不认为有任何理由将上述活动分成单独的 CPU 线程。这不必要地使事情复杂化。管理单个设备最无故障的方法是使用单个 CPU 线程。