cuda - CUDA 中的同步

Question

我阅读了有关 cuda 同步的 cuda 参考手册，但我不清楚。例如为什么我们使用cudaDeviceSynchronize()or __syncthreads()？如果不使用它们会发生什么，程序无法正常工作？cudaMemcpy和cudaMemcpyAsync在行动有什么区别？
你能举一个例子来说明这种差异吗？

Answer 1

cudaDeviceSynchronize()当希望 CPU 活动等待任何挂起的 GPU 活动完成时，在主机代码（即在 CPU 上运行）中使用。在许多情况下，没有必要显式地执行此操作，因为发布到单个流的 GPU 操作会自动序列化，并且某些其他操作（例如cudaMemcpy()，它们内置了固有的阻塞设备同步）。但是对于其他一些目的，例如调试代码，强制设备完成任何未完成的活动可能很方便。

__syncthreads()在设备代码中使用（即在 GPU 上运行），在具有独立并行操作的代码中可能根本不需要（例如将两个向量逐个元素相加）。然而，一个常用的例子是在共享内存之外运行的算法。在这些情况下，经常需要将全局内存中的值加载到共享内存中，并且我们希望线程块中的每个线程都有机会在任何实际处理发生之前加载其适当的共享内存位置。在这种情况下，我们要__syncthreads()在处理发生之前使用，以确保完全填充共享内存。这只是一个例子。 __syncthreads()可以在需要在线程块内进行同步的任何时候使用。它不允许块之间的同步。

cudaMemcpy和的区别在于cudaMemcpyAsync，非异步版本的调用只能发布到流 0，并且会阻塞调用 CPU 线程，直到复制完成。异步版本可以选择采用流参数，并在复制完成之前立即将控制权返回给调用线程。异步版本通常在我们想要异步并发执行的情况下使用。

如果您有关于 CUDA 编程的基本问题，建议您参加一些可用的网络研讨会。

Answer 2

此外，__syncthreads()当您的代码中有一些条件路径，然后您想要运行依赖于多个数组元素的操作时，它变得非常必要。考虑以下示例：

int n = threadIdx.x;

if( myarray[n] > 0 )
{
     myarray[n] = - myarray[n];
}
double y = myarray[n] + myarray[n+1]; // Not all threads reaches here at the same time

在上面的示例中，并非所有线程都有相同的执行顺序。根据if条件，某些线程将花费更长的时间。在考虑示例的最后一行时，您需要确保所有线程都完全完成了 if 条件并myarray正确更新。如果不是这种情况，y可以使用一些更新和未更新的值。在这种情况下，必须__syncthreads()在评估之前添加y以克服此问题：

if( myarray[n] > 0 )
{
     myarray[n] = - myarray[n];
}
__syncthreads(); // All threads will wait till they come to this point
// We are now quite confident that all array values are updated.
double y = myarray[n] + myarray[n+1];