cuda - 在 cuda 中，哪个线程负责内核内请求的共享内存分配？

Question

我了解内核中的指令由所有线程执行。让我们考虑以下情况：

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

基本上，这段代码将作为线程在不同的内核上运行。
现在有一个共享内存分配。既然所有线程都会遇到它（共享内存分配），那么它会被所有线程分配吗？（逻辑上不是。）但我确信至少有一个线程必须分配它。我想知道是哪个线程做的？请帮助我了解我的理解是错误的。

标签： cuda

Answer 1

没有线程负责此分配 - 线程不运行任何涉及分配此内存的 SASS 代码。

如果您使用动态 ( extern) 共享分配，则同样的陈述是正确的——没有线程负责——这意味着线程不运行任何涉及分配此内存的 SASS 代码。不涉及函数调用或其他机制。

当线程 SASS 代码（即内核）开始执行时，内存已经分配，​​并且已经建立了指向它的指针。

有一个皱纹需要注意。如果共享内存声明涉及构造函数，则构造函数将在所有线程上运行。这可能是令人困惑的行为。