cuda 将大于线程数的数组复制到共享内存

Question

如果内核出于某些其他原因需要 y 线程，那么有人会如何复制大小为 x 的数组。我找到的所有示例都是复制线程数大小的数组，例如：

__global__ void staticReverse(int *d, int n)
{
  __shared__ int s[64];
  int t = threadIdx.x;

  s[t] = d[t];
  __syncthreads();

如果我需要我的 s[100000] 并且我需要仅用 1 个 640 线程块启动我的内核，它会是什么样子？
如果我需要我的 s[100000] 并且我需要用 10 个 64 线程块启动我的内核，它会是什么样子？在这里我也很困惑，因为共享内存只是1个块，所以我不明白其他块的线程如何复制到哪个块的共享内存？

Answer 1

How would it look If I need my s[100000] and I need to launch my kernel with just 1 block of 640 threads?

不会。当前支持的 CUDA 硬件 limit 每块 48kb 或 96kb，因此具有 100000 个元素的整数共享数组是非法的。

但一般来说，如果您的设计模式需要将每个线程的多个数据点加载到共享内存中，那么您会这样做

__global__ void staticReverse(int *d, int n)
{
  __shared__ int s[2048];
  int t = threadIdx.x;

  for(; t < 2048; t += blockDim.x) s[t] = d[t];
  __syncthreads();

因为共享内存是严格的块范围，所以您必须使用类似上述设计模式的东西来将数据加载到共享存储。显然，来自其他块的线程永远无法访问除它们自己的块之外的任何共享内存。

cuda 将大于线程数的数组复制到共享内存

cuda copy to shared memory an array larger than thread count

c++

cuda

shared-memory