cuda 将数组从全局内存复制到共享内存答案

【问题标题】：cuda copy array from global memory to shared memorycuda 将数组从全局内存复制到共享内存
【发布时间】：2013-01-11 22:15:08
【问题描述】：

hy，我正在尝试将 2 个数组从全局内存复制到共享内存：全局数组类型是双精度的，每个人都有 32 个元素。网格是一维的，块是一维的。网格维度为 10000，NumberThreadPerBlock 为 32

__global__ void kernel_0(double px[], double py[], int N)
{
int ii,

    jj,tid;
    tid=blockIdx.x*blockDim.x + threadIdx.x;
    __shared__ double s_px[256];
    __shared__ double s_py[256];
    __shared__ double s[256];

s_px[threadIdx.x]=px[tid];
s_py[threadIdx.x]=py[tid];
s[threadIdx.x]=py[tid];
__syncthreads();
}


int main (int argc, char *argv[]){
    double *px, *py , *x, *y, PI, step, *d_x, *d_y,*d_px, *d_py,sharedMemSize;
    int N, Nx, ii;
    PI = 4*atan(1.0);
    Nx = 10000; 
    N = 32; 

    px = (double *) malloc(N*sizeof(double));
    py = (double *) malloc(N*sizeof(double));

    // lookup table: sin // from 0 to PI 
    step = 1.0 / (N-1);
    for (ii = 0; ii < N; ii++){ 
        px[ii] = ii*step*PI;
        py[ii] = sin(px[ii]);
    }   

    cudaMalloc( (void **) &d_px, N*sizeof(double) );
    cudaMalloc( (void **) &d_py, N*sizeof(double) );        

    cudaMemcpy( d_px, px, N*sizeof(double), cudaMemcpyHostToDevice );
    cudaMemcpy( d_py, py, N*sizeof(double), cudaMemcpyHostToDevice );

    dim3 dimGrid(Nx);
    dim3 dimBlock(N,1,1);
    kernel_0<<< dimGrid, dimBlock>>>(px, py, N);

}

它可以编译，但 cuda-memmcheck 显示很多错误：

========= Invalid __global__ read of size 8
=========     at 0x00000058 in kernel_0
=========     by thread (31,0,0) in block (6,0,0)
=========     Address 0x11e0db38 is out of bounds
=========
========= ERROR SUMMARY: 96 errors

你能帮帮我吗？

【问题讨论】：

这个内核不会像这里显示的那样运行。请发布您正在使用的实际内核。
还请包括您的全局数组分配和内核调用。如果您发布一个完整、简单、可编译的示例来演示该问题，这可能是最好的。
我敢打赌，tid 被用于px 或py 的读取之一，但您必须显示您的实际代码确认。

标签： cuda shared-memory

【解决方案1】：

据我所知，分配给设备指针 (px, py) 的内存是 32*sizeof(double)，但是您拥有的块数是 10000。设备内存是全局的，所有块共享它，只有共享内存是为每个块定义的。因此，对于 blockId.x = 1，您应该获得无效的内存访问。而且，在内核调用中，应该是d_px，d_py。

【讨论】：