CudaDeviceSynchronize vs cudaThreadSynchronize vs cudaStreamSynchronize

首先对这三个函数做一下解释：

cudaDeviceSynchronize() 会阻塞当前程序的执行，直到所有任务都处理完毕（这里的任务其实就是指的是所有的线程都已经执行完了kernel function）。

cudaThreadSynchronize()的功能和cudaDeviceSynchronize()基本上一样，这个函数在新版本的cuda中已经被“废弃”了，不推荐使用，如果程序中真的需要做同步操作，推荐使用cudaDeviceSynchronize()。

cudaStreamSynchronize()和上面的两个函数类似，这个函数带有一个参数，cuda流ID，它只阻塞那些cuda流ID等于参数中指定ID的那些cuda例程，对于那些流ID不等的例程，还是异步执行的。

按照Cuda官方文档上的解释，cuda kernel函数是异步执行的，也就是说，kernel函数在调用之后立即把控制权交换给CPU，CPU接着往下执行。基于这样的解释，那么我们在编写cuda程序时，比如：

kernel1<<<X,Y>>>(...);
kernel2<<<X,Y>>>(...);
cudaMemcpy(...);

是否应该在每一个kernel function之后加上同步的语句呢?或者说，什么时候我们应该加上同步语句呢？

其实，尽管kernel function是以异步的方式执行的，但对于同一个流内的所有cuda例程却是顺序执行的，这在官方文档上也有说明。而在我们的cuda例程中如果不指定流ID，流ID采用缺省值，也就是说，诸如上面的代码，不加同步语句也是可以的。当然，如果你还是不放心，加上同步语句也无所谓，无非可能需要多耗费一点同步的时间而已。

但是，当我们的程序中有多个流，并且流之间在某一点需要通信时，那就必须在这一点处加上同步的语句，即cudaDeviceSynchronize()。

另外说明一点，cudaMemcpy函数对于主机是同步的，但是cudaMemcpyAsyn是异步的，所以，使用 cudaMemcpyAsyn()后加  cudaDevicesynchronize() ，与单一条cudaMemcpy()，效果上是相同的。