CUDA Streams

https://www.cnblogs.com/shrimp-can/p/5231857.html

转载请注明来源：http://www.cnblogs.com/shrimp-can/p/5231857.html

简单可以理解为：cudaMemcpy是同步的，而cudaMemcpyAsync是异步的。具体理解需要弄清以下概念：

1.CUDA Streams

在cuda中一个Stream是由主机代码发布的一系列再设备上执行的操作，必须确保顺序执行。不同streams里面的操作可以交叉执行或者并发执行。

2.默认stream

设备操作包括：数据传输和kernels，在cuda中，所有的设备操作都在stream中执行。当没有指定stream时，使用默认的stream。默认stream是一个针对设备操作同步的stream，也就是说，只有当所有之前设备上任何stream里面的操作全部完成时，才开始默认stream里面操作的执行，并且默认stream里面的一个操作必须完成，其他任何stream里面的操作才能开始。

例如以下代码：

cudaMemcpy(d_a, a, numBytes, cudaMemcpyHostToDevice);
increment<<<1,N>>>(d_a)
cudaMemcpy(a, d_a, numBytes, cudaMemcpyDeviceToHost);

从设备端来看，这三个操作都在默认stream中，并且按顺序执行；从主机端来看，数据传输是阻塞的或者同步传输，而kernel是异步的。第一步主机到设备的数据传输是同步的，CPU线程不能到达第二行直到主机到设备的数据传输完成。一旦kernel被处理，CPU线程移到第三行，但是改行的传输不能开始，因为设备端正在执行第二行的内容。

3.非默认stream

非默认stream中的数据传输使用函数cudaMemcpyAsync()，这个函数在主机端是非阻塞的，传输处理后控制权马上返回给主机线程

参考：https://devblogs.nvidia.com/parallelforall/how-overlap-data-transfers-cuda-cc/