Re: [問題] CUDA 矩陣相加 error

[lgen7604 ()]

我要先提醒一件事

在CUDA中是沒有辦法使用 2D array 的

雖然CUDA中有 struct cudaArray 的架構

不過是專門給 texture 用的

必須使用 texture fetch 的方式來讀取(而且是唯讀)

詳細內容可以參考 CUDA Programming Guide

一般來說處理 2D array 運算最常用的方法

就是將 2D array 視為 1D array 來處理

照這個想法來看的話

程式可以做以下的修改

※ 引述《aada (aada)》之銘言：
: 大家好, 想請教個問題 以下是我的程式,主要是CUDA運算矩陣加法
: 有些error不曉得要從拿邊修改起,
: #include<stdio.h>
: #include<stdlib.h>
: #include<time.h>
: #include<cuda.h>
: __global__ void VecAdd(int* A[2][2], int* B[2][2], int* C[2][2])
                         int* A, int* B, int* C, int width
原本使用pointer的2D array感覺有點怪
可以用pointer指向1D array,再加上寬度的參數來模擬2D的處理
: {
:         int i = blockIdx.x*blockDim.x+threadIdx.x; //用多區塊,多執行緒的寫法
          int col = blockIdx.x*blockDim.x+threadIdx.x; // x 表示 column
:         int j = blockIdx.y*blockDim.y+threadIdx.y; //用多區塊,多執行緒的寫法
          int row = blockIdx.y*blockDim.y+threadIdx.y; // y 表示 row
:         C[i][j] = A[i][j]+B[i][j];
          C[row*width+col] = A[row*width+col] + B[row*width+col];
: }
: int main()
: {
:         int N = 2;
:         int h_A[2][2]={{1,2},{3,4}};
:         int h_B[2][2]={{5,6},{7,8}};
:         int *h_B2[2][2];
          int h_B2[2][2];
:         int size = N*sizeof(int);
          int size = N*N*sizeof(int); // 應該是N*N吧
:         int Grid = 1;
:         int Block = 10;
          dim3 Block(N,N);
在決定 Grid 和 Block 的 dimension 的時候
通常是依據運算的內容
因為你現在處理的是很小的 2D array
所以可以只用1個Block就處理完
而thread我設為N*N (為了配合N*N array的加法運算)
:         int* d_A;
:         cudaMalloc((void**)&d_A, size);
:         int* d_B;
:         cudaMalloc((void**)&d_B, size);
:         int* d_C;
:         cudaMalloc((void**)&d_C, size);
:         cudaMemcpy(d_A, h_A, size, cudaMemcpyHostToDevice);
:         cudaMemcpy(d_B, h_B, size, cudaMemcpyHostToDevice);
:         VecAdd<<<Grid, Block>>>(d_A, d_B, d_C);
          VecAdd<<<Grid, Block>>>(d_A, d_B, d_C, N);
:         cudaMemcpy(h_B2, d_C, size, cudaMemcpyDeviceToHost);
:         for(int i=0;i<N;++i)
:         {
:                 for(int j=0;j<N;++j)
:                 {
:                         printf("%d+%d=%d\n",h_A[i][j],h_B[i][j],h_B2[i][j]);
:                 }
:         }
:         cudaFree(d_A);
:         cudaFree(d_B);
:         cudaFree(d_C);
:         system("PAUSE");
:         return 0;
: }
: matrix_add.cu(11): error: expression must have integral or enum type
: matrix_add.cu(43): error: argument of type "int *" is incompatible with
: parameter of type "int *(*)[2]"
: matrix_add.cu(43): error: argument of type "int *" is incompatible with
: parameter of type "int *(*)[2]"
: matrix_add.cu(43): error: argument of type "int *" is incompatible with
: parameter of type "int *(*)[2]"
: 應該要怎麼改呢,謝謝

大致上這樣改就可以work了

不過當array很大  無法用1個block處理的時候

該如何分配每個block內的每個thread的工作

可以想想看

另外關於 memory access 的 performance 問題....

太複雜懶得提了  orz

先把 2D array 以 1D 形式處理的基本搞懂

其他的再慢慢說吧

--
※ 發信站: 批踢踢實業坊(ptt.cc) 
◆ From: 122.120.33.65
推 aada:謝謝L大的幫忙,
推 andyjy12:2D array 會有讀取對齊的問題，可以去看一下malloc2D (?)
→ lgen7604:memory讀取有alignment和bank conflict的問題  不過太麻
→ lgen7604:煩懶得提 詳細內容一樣請參考 CUDA Programming Guide
→ lgen7604:還有 要處理array的alignment問題  可以使用
→ lgen7604:cudaMallocPitch和cudaMalloc3D  很遺憾沒有malloc2D
→ lgen7604:不過malloc的結果還是 linear memory 就是了
推 andyjy12:sorry~記錯function