CUDA线程同步

Question

我对同步有点困惑。

1. 使用__syncthreads，您可以同步block.This中的线程，(使用__syncthreads)必须只使用共享内存吗？还是在__syncthreads中使用共享内存具有最佳的性能？
2. 通常，只有当线程存在于同一个线程块中时，线程才能安全地相互通信，对吗？那么，为什么我们不总是使用共享内存呢？因为它不够大？如果我们不使用共享内存，我们如何确保结果是正确的？
3. 我有一个程序，有时运行ok (我得到结果)，有时我得到'nan‘的结果，而不改变任何东西。这会是一个同步问题吗？

Answer 1

1. __syncthreads的使用不涉及共享内存，它只确保块内的同步。但是，当您希望线程通过共享内存共享数据时，需要同步它们。
2. 我们并不总是使用共享内存，因为它很小，而且当应用程序使用不好时，它会减慢应用程序的速度。这是由于潜在的银行冲突时，错误寻址共享内存。此外，最近的体系结构(从2.0)实现共享内存在相同的硬件区域，而不是缓存。因此，一些经验丰富的CUDA开发人员建议不要使用共享内存，而只依赖缓存机制。
3. 是可以的。如果您想知道这是否是死锁，请尝试增加所使用的块的数量。如果是死锁，您的GPU应该会冻结。如果不是，发布您的代码，我们将更容易回答;)

Answer 2

__syncthreads()和共享内存是独立的想法，您不需要一个来使用另一个。我想到的使用__syncthreads()的唯一要求是，所有线程都必须最终到达代码中的点，否则程序就会挂起。

至于共享内存，是的，这可能是一个大小的问题，您没有看到它一直在使用。根据我的理解，共享内存被分割在所有块中。例如，使用共享内存1kb和100个块启动内核需要100 1kb，这超出了SM上可用的范围。

Answer 3

1. 虽然shared memory和__syncthreads()是独立的概念，但它们往往是相辅相成的。否则，如果线程独立运行，则不需要使用__syncthreads()。
2. 有两个方面限制了shared memory的使用: 1)。shared memory的大小是有限的。要获得最佳性能，您需要在使用shared memory时避免银行冲突。
3. 这可能是由于缺乏__syncthreads()。有时，在不使用shared memory的情况下使用__syncthreads()可能会导致不可预测的结果。