在cuda中静态分配共享内存的正确方法是什么?为什么

Question

在以下两种静态分配共享内存的方法中，哪种方法是正确的，为什么？我得到了相同的结果，但我正在尝试更详细地了解行为。

内核1:

__shared__ int as[3][3],bs[3][3];
__global__ void Sharesum(int* a,int* b,int* c,int n)
{
int s,k,i,sum=0;
int tx,ty,bx,by;
tx=threadIdx.x;
ty=threadIdx.y;
as[ty][tx]=a[tx+n*ty];
bs[ty][tx]=b[tx+n*ty];
sum += as[ty][tx]+bs[ty][tx];
c[tx*n+ty]=sum;
}

内核2:

__global__ void Sharesum(int* a,int* b,int* c,int n)
{
__shared__ int as[3][3],bs[3][3];
int s,k,i,sum=0;
int tx,ty,bx,by;
tx=threadIdx.x;
ty=threadIdx.y;
as[ty][tx]=a[tx+n*ty];
bs[ty][tx]=b[tx+n*ty];
sum += as[ty][tx]+bs[ty][tx];
c[tx*n+ty]=sum;
}

Answer 1

对于所显示的内容，这两种方法之间不应该有任何区别。我不确定是否有一个答案表明；正确的"；其中一个不是。

然而，我们可以称之为"；全球范围"；声明，影响模块中定义的所有内核。这意味着所有内核都将根据全局定义保留并具有可用的共享分配。

第二个只影响它所作用域的内核

其中一个或两个都可能是正确的，这取决于你想要的意图。