如何在CUDA中使用二维数组？

Question

如何分配一个大小为MXN的二维数组？以及如何在CUDA中遍历该数组？

__global__ void test(int A[BLOCK_SIZE][BLOCK_SIZE], int B[BLOCK_SIZE][BLOCK_SIZE],int C[BLOCK_SIZE][BLOCK_SIZE])
{

    int i = blockIdx.y * blockDim.y + threadIdx.y;
    int j = blockIdx.x * blockDim.x + threadIdx.x;
    
    if (i < BLOCK_SIZE && j < BLOCK_SIZE)
        C[i][j] = A[i][j] + B[i][j];

}

int main()
{

    int d_A[BLOCK_SIZE][BLOCK_SIZE];
    int d_B[BLOCK_SIZE][BLOCK_SIZE];
    int d_C[BLOCK_SIZE][BLOCK_SIZE];

    int C[BLOCK_SIZE][BLOCK_SIZE];

    for(int i=0;i<BLOCK_SIZE;i++)
      for(int j=0;j<BLOCK_SIZE;j++)
      {
        d_A[i][j]=i+j;
        d_B[i][j]=i+j;
      }
    

    dim3 dimBlock(BLOCK_SIZE, BLOCK_SIZE); 
    dim3 dimGrid(GRID_SIZE, GRID_SIZE); 

    test<<<dimGrid, dimBlock>>>(d_A,d_B,d_C); 

    cudaMemcpy(C,d_C,BLOCK_SIZE*BLOCK_SIZE , cudaMemcpyDeviceToHost);

    for(int i=0;i<BLOCK_SIZE;i++)
      for(int j=0;j<BLOCK_SIZE;j++)
      {
        printf("%d\n",C[i][j]);
    
      }
}

- Sandeep

---

你不能使用cudaMemcpy获取2D数组的值，相反，你必须像@Dave所说的那样使用cudaMallocPitch或cudaPitchPtr与cudaMalloc3D一起使用。 - ardiyu07

Answer 1

如何分配二维数组：

int main() {
    #define BLOCK_SIZE 16
    #define GRID_SIZE 1
    int d_A[BLOCK_SIZE][BLOCK_SIZE];
    int d_B[BLOCK_SIZE][BLOCK_SIZE];

    /* d_A initialization */

    dim3 dimBlock(BLOCK_SIZE, BLOCK_SIZE); // so your threads are BLOCK_SIZE*BLOCK_SIZE, 256 in this case
    dim3 dimGrid(GRID_SIZE, GRID_SIZE); // 1*1 blocks in a grid
    
    YourKernel<<<dimGrid, dimBlock>>>(d_A,d_B); //Kernel invocation
}

如何遍历该数组：

__global__ void YourKernel(int d_A[BLOCK_SIZE][BLOCK_SIZE], int d_B[BLOCK_SIZE][BLOCK_SIZE]){
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int col = blockIdx.x * blockDim.x + threadIdx.x;
    if (row >= h || col >= w) return;
    /* whatever you wanna do with d_A[][] and d_B[][] */
}

我希望这对你有帮助，你还可以参考关于矩阵乘法的CUDA编程指南

- ardiyu07

---

虽然这种方法可行，但它只是在设备内存中创建了一个巨大的线性数组。您还可以使用cudaMalloc3D分配二维数组，这些数组针对二维数据访问进行了优化。我不知道您只是想要二维数组的索引还是性能。 - Dave O.

@username_4567，这就是/* d_A初始化 */的意思。然而，内存释放却没有实现。 - Denys S.

/* d_A initialization */ 的实际内容也是答案的重要部分。您能否提供它？ - JRsz

Cuda内核不使用return。 - user14518353

Answer 2

最好的方法是以向量形式存储二维数组A。 例如，您有一个大小为nxm的矩阵A，其在指针表示法中的（i，j）元素将为

A[i][j] (with i=0..n-1 and j=0..m-1). 

你可以用向量形式写成

A[i*n+j] (with i=0..n-1 and j=0..m-1).

在这种情况下使用一维数组将简化复制过程，这将是简单的：

double *A,*dev_A; //A-hous pointer, dev_A - device pointer;
A=(double*)malloc(n*m*sizeof(double));
cudaMalloc((void**)&dev_A,n*m*sizeof(double));
cudaMemcpy(&dev_A,&A,n*m*sizeof(double),cudaMemcpyHostToDevice); //In case if A is double

- Suren