c++ - 在 CUDA 中共享内存？ CODE 是如何工作的？

Question

我有一个程序来计算数组的值:数组 A :有 32 个元素，值形式 0 -> 31。数组 B :有 16 个元素，值 = 0；

**我想按照这个规则计算 B[i] 的值:B[i]=A[i*2] + A[i*2+1];我从 0 到 31 ** 我在我的示例代码中使用 CUDA 编程:

主.cu

 __global__ void Kernel(int *devB, int *devA)
    {  
    // Use share memory, 16 thread per block, so I use 16element for share memory in block 
    __shared__ int smA[16];
    //copy data from global memory to shared memory
    //1 thread copies 1 elementente
      smA[threadIdx.x] = devA[threadIdx.x + blockIdx.x * blockDim.x];
      __syncthreads(); 
    //8 thread in Block
      if (threadIdx.x < 8) 
    {
      devB[threadIdx.x + blockIdx.x * blockDim.x] = 
      smA[threadIdx.x * 2] + smA[threadIdx.x * 2 + 1];
    }
}

虚空主体

void main()
{
int *A = (int*)malloc(sizeof(int) * 32);
int *B = (int*)malloc(sizeof(int) * 16);

for (int i = 0; i < 32; i++)
A[i] = i;


    int *devA = NULL;
    cudaMalloc((void**)&devA, sizeof(int) * 32);
    cudaMemcpy(devA, A, sizeof(int) * 32), cudaMemcpyHostToDevice);
    int * devB = NULL;
    cudaMalloc((void**)&devB, sizeof(int) * 16);

    dim3 block(16, 1, 1);
    dim3 grid(2, 1, 1);

    Kernel<<<grid, block>>>(devB, devA);

    //copy back data to host
    cudaMemcpy(B, devB, sizeof(int) * 16, cudaMemcpyDeviceToHost);
    for (int i = 0; i < 16; i++) printf("%d/t", b[i]);

    if (A != NULL) free(A);
    if (B != NULL) free(B);
    if (devA != NULL) cudaFree(devA);
    if (devB != NULL) cudaFree(devB); 
}

所以，我想问一下:按照我上面的代码，我在内核中使用了Share memory int smnA[16]，并且2 block = 2*16 thread 因为每个线程执行一个内核(来自Seland.pdf)=> 我将在共享内存中有 16x16 = 256 个元素 ？ => 它没有逻辑!

Answer 1

不，你的假设是错误的。因为共享内存可用于同一 block 内的线程交互，所以共享内存也分配给整个线程 block 。在您的示例中，您将为每个线程 block 使用 16 个整数元素。您的内核总共需要 32 个整数元素才能同时运行所有线程 block 。即使它不一样，但也许您可以将它与 C 代码中的静态变量进行比较。

如果您在内核中编写类似于以下代码示例的代码，每个线程都将使用它自己的包含 16 个元素的数组。但是这个数组不能被其他线程访问(洗牌指令除外)。

__globa__ void kernel (...)
{
  int array_single_thread[16]; // Every thread instance has it's own array.
  ...
  __shared__ int array_thread_block[16]; // Once allocated for complete thread block.
}