cuda 如何处理内核中的 __syncthreads()？答案

【问题标题】：How cuda handle __syncthreads() in kernel?cuda 如何处理内核中的 __syncthreads()？
【发布时间】：2017-02-27 13:37:52
【问题描述】：

认为我有一个大小为 1024 的块，并假设我的 gpu 有 192 个 cuda 核心。当 cuda 内核大小小于块大小时，cuda 如何处理内核中的 __syncthreads()？

__global__ void staticReverse(int *d, int n)
{
  __shared__ int s[1024];
  int t = threadIdx.x;
  int tr = n-t-1;
  s[t] = d[t];
  __syncthreads();
  d[t] = s[tr];
}

“tr”在本地内存中的剩余情况如何？

【问题讨论】：

标签： cuda

【解决方案1】：

我认为你混合了一些东西。

首先，具有 192 个 CUDA 核心的 GPU 是总核心数。然而，每个块都映射到单个 流式多处理器 (SM)，它可能具有较低的核心数（取决于 GPU 代）。

假设您拥有一个 Pascal GPU，每个 SM 有 64 个内核，并且您有 3 个短信。单个块映射到单个 SM。因此，您将有 64 个内核同时处理 1024 个线程。这样的 SM 有足够的寄存器来保存 1024 个线程的所有必要数据，但它只有 64 个内核，可以快速交换它们正在处理的线程。

这样所有本地数据，例如tr 可以保留在内存中。

现在，由于这种快速交换和并发执行，可能会发生——完全是偶然的——一些线程领先于其他线程。如果你想确保在某个时刻所有线程都在同一个位置，你可以使用__syncthreads()。该函数所做的只是指示调度程序正确地将工作分配给 CUDA 内核，以便它们在某个时刻都处于程序中的那个位置。

【讨论】：