Cuda矩陣乘法結果不同於MATLAB

Question

它已經兩天了，我仍然無法弄清楚爲什麼我的CUDA矩陣乘法的實現不同於MATLAB中產生的結果。

CUDA內核：A（200x60000）= W（200x784）*數據（784x6000）

__global__ void CalculateA(Matrix W, Matrix Data, Matrix A) 
{ 
    int Row = blockIdx.y * blockDim.y + threadIdx.y; 
    int Col = blockIdx.x * blockDim.x + threadIdx.x; 
    if ((Row < W.row) && (Col < Data.col)){ 
     float Cvalue = 0.0; 
     for (int i = 0; i < W.col; ++i){ 
      Cvalue += W.elements[Row*W.col+i] * Data.elements[i*Data.col+Col]; 
     } 
    A.elements[Row*A.col+Col] = Cvalue; 
    } 
} 

並調用內核：

void myFunc(Matrix W1, Matrix data){ 
     Matrix d_W1, d_data, d_a2, a2; 
    size_t size; 

    a2.row = W1.row; d_a2.row = a2.row; 
    a2.col = data.col; d_a2.col = a2.col; 
    size = a2.col*a2.row*sizeof(float); 
    cudaMalloc(&d_a2.elements,size); 

    d_W1.row = W1.row; d_W1.col = W1.col; 
    size = W1.col*W1.row*sizeof(float); 
    cudaMalloc(&d_W1.elements,size); 
    cudaMemcpy(d_W1.elements,W1.elements,size,cudaMemcpyHostToDevice); 

    d_data.col = data.col; d_data.row = data.row; 
    size = data.row*data.col*sizeof(float); 
    cudaMalloc(&d_data.elements,size); 
    cudaMemcpy(d_data.elements,data.elements,size,cudaMemcpyHostToDevice); 
    dim3 dimGrid(data.col/32 + 1, W1.row/32 + 1, 1); 
    dim3 dimBlock(32, 32, 1); 

    CalculateA<<<dimGrid, dimBlock>>>(d_W1, d_data, d_a2); 
    a2.elements = new float [a2.row*a2.col]; 
    cudaMemcpy(a2.elements,d_a2.elements,sizeof(float)*a2.row*a2.col,cudaMemcpyDeviceToHost); 

    printf("\nA2 first and last member %f - %f\n",a2.elements[0],a2.elements[a2.row*a2.col-1]); 
} 

結果差不低例如第一和最後CUDA代碼的元素爲0.011322和-0.179534，但在MATLAB中乘以0.4280和0.0056。

這是我要做的事在MATLAB：

>> size(W1)  ans =  200 784 

>> size(data) ans =  784  60000 

>> z2=W1*data; 

>> size(z2)  ans =  200  60000 

>> z2 = z2(:); 

>> z2(1)  ans = 0.4280 

>> z2(200*60000)ans = 0.0056 

Answer 1

沒有什麼錯，你發佈的代碼。如果我擴大你的內核和功能成爲一個完整的運行例子是這樣的：

#include <iostream> 

struct Matrix 
{ 
    int row; 
    int col; 
    float *elements; 

    __device__ __host__ 
    float& operator()(int r, int c) { return elements[r*col + c]; }; 
}; 

__global__ void CalculateA(Matrix W, Matrix Data, Matrix A) 
{ 
    int Row = blockIdx.y * blockDim.y + threadIdx.y; 
    int Col = blockIdx.x * blockDim.x + threadIdx.x; 
    if ((Row < W.row) && (Col < Data.col)){ 
     float Cvalue = 0.0; 
     for (int i = 0; i < W.col; ++i){ 
      Cvalue += W.elements[Row*W.col+i] * Data.elements[i*Data.col+Col]; 
     } 
    A.elements[Row*A.col+Col] = Cvalue; 
    } 
} 

void myFunc(Matrix W1, Matrix data) 
{ 
    Matrix d_W1, d_data, d_a2, a2; 
    size_t size; 

    a2.row = W1.row; d_a2.row = a2.row; 
    a2.col = data.col; d_a2.col = a2.col; 
    size = a2.col*a2.row*sizeof(float); 
    cudaMalloc(&d_a2.elements,size); 

    d_W1.row = W1.row; d_W1.col = W1.col; 
    size = W1.col*W1.row*sizeof(float); 
    cudaMalloc(&d_W1.elements,size); 
    cudaMemcpy(d_W1.elements,W1.elements,size,cudaMemcpyHostToDevice); 

    d_data.col = data.col; d_data.row = data.row; 
    size = data.row*data.col*sizeof(float); 
    cudaMalloc(&d_data.elements,size); 
    cudaMemcpy(d_data.elements,data.elements,size,cudaMemcpyHostToDevice); 
    dim3 dimGrid(data.col/32 + 1, W1.row/32 + 1, 1); 
    dim3 dimBlock(32, 32, 1); 

    CalculateA<<<dimGrid, dimBlock>>>(d_W1, d_data, d_a2); 
    a2.elements = new float [a2.row*a2.col]; 
    cudaMemcpy(a2.elements,d_a2.elements,sizeof(float)*a2.row*a2.col,cudaMemcpyDeviceToHost); 

    for(int j=0; j<a2.col; ++j) { 
     for(int i=0; i<a2.row; ++i) { 
      std::cout << a2(i,j) << " "; 
     } 
     std::cout << std::endl; 
    } 
} 

int main(void) 
{ 
    float a[6] = { 1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f }; 
    float b[6] = { 0.1f, 0.2f, 0.3f, 0.4f, 0.5f, 0.6f}; 

    Matrix W1; W1.row=2; W1.col=3; W1.elements = &a[0]; 
    Matrix Data; Data.row=3; Data.col=2; Data.elements = &b[0]; 

    myFunc(W1, Data); 

    return 0; 
} 

，並運行它，我得到這個：

>nvcc -arch=sm_21 -Xptxas="-v" -m32 matrix.cu 
matrix.cu 
tmpxft_000014f4_00000000-5_matrix.cudafe1.gpu 
tmpxft_000014f4_00000000-10_matrix.cudafe2.gpu 
matrix.cu 
ptxas : info : 132 bytes gmem, 28 bytes cmem[14] 
ptxas : info : Compiling entry function '_Z10CalculateA6MatrixS_S_' for 'sm_21' 
ptxas : info : Function properties for _Z10CalculateA6MatrixS_S_ 
    0 bytes stack frame, 0 bytes spill stores, 0 bytes spill loads 
ptxas : info : Used 14 registers, 68 bytes cmem[0] 
tmpxft_000014f4_00000000-5_matrix.cudafe1.cpp 
tmpxft_000014f4_00000000-15_matrix.ii 

>cuda-memcheck a.exe 
========= CUDA-MEMCHECK 
2.2 4.9 
2.8 6.4 
========= ERROR SUMMARY: 0 errors 

這是點產品的正確答案假設列優先順序（這是Matlab慣例）。

所以如果你的結果不一致，那是因爲你沒有向我們展示過的東西。一種可能是你的測試問題太大（而且內核如此低效），如果你在顯示GPU上運行這個問題，你的程序將觸發顯示驅動看門狗定時器的限制，並在內核結束運行之前被殺死。另外請注意，你有沒有 CUDA API錯誤檢查什麼，所以它可能會得到運行時錯誤，它可能會阻止你的內核完成或甚至運行，但你根本沒有注意到，因爲缺乏錯誤檢查。