2013-01-21 78 views
2

我在CUDA中編寫了簡單的平鋪矩陣乘法。它是這樣的:OpenCL和CUDA之間的性能差距非常明顯

__global__ void matrixMultiplyShared(float * A, float * B, float * C, 
         int numARows, int numAColumns, 
         int numBRows, int numBColumns, 
         int numCRows, int numCColumns) { 

    __shared__ float ds_A[TILE_WIDTH][TILE_WIDTH]; 
    __shared__ float ds_B[TILE_WIDTH][TILE_WIDTH]; 

    int bx = blockIdx.x; int by = blockIdx.y; 
    int tx = threadIdx.x; int ty = threadIdx.y; 

    int row = by * TILE_WIDTH + ty; 
    int col = bx * TILE_WIDTH + tx; 

    float Cvalue = 0.0; 

// Loop over the M and N tiles required to compute the Pd element 
    for (int m = 0; m < (numAColumns-1)/TILE_WIDTH+1; ++m) { 
     if(row<numARows && m*TILE_WIDTH+tx < numAColumns){ 
      ds_A[ty][tx] = A[row*numAColumns + m*TILE_WIDTH+tx]; 
     } else { 
      ds_A[ty][tx] = 0; 
     } 
     if(m*TILE_WIDTH+ty < numBRows && col < numBColumns){ 
      ds_B[ty][tx] = B[(m*TILE_WIDTH+ty)*numBColumns+col]; 
     } else { 
      ds_B[ty][tx] = 0; 
     } 
     __syncthreads(); 
     if(row < numCRows && col < numCColumns){ 
      for (int k = 0; k < TILE_WIDTH; ++k) 
       Cvalue += ds_A[ty][k] * ds_B[k][tx]; 
     } 
     __syncthreads(); 
    } 
    if(row < numCRows && col < numCColumns) 
     C[row*numCColumns+col] = Cvalue; 
} 

在那之後,我用同樣的上述內核(有一些微小的變化)中的OpenCL版本CUDA和OpenCL性能放在一起比較。但結果遠遠超出了我的預期。 OpenCL比CUDA快6-7倍。它有效嗎? Nisght的輸出如下:

CUDA: CUDA Nisght output: Kernel Ex time: 3.78s

的OpenCL: CUDA Nisght output: Kernel Ex time: 0.53s

你可以看到啓動應用程序和執行內核之間有很大的差距。爲什麼會發生?


我的GPU是:GTX 580 |內核Ex時間(CUDA):3.78s |內核Ex時間(OpenCL):0.53s |

CUDA代碼:http://pastebin.com/VQMp3Hba

OpenCL的主機代碼:http://pastebin.com/cjGYSLQf

的OpenCL內核代碼:http://pastebin.com/KKw3Ayz7

+3

我想你應該爲OpenCL實現添加代碼,那麼至少有機會找到差異的原因... – ppeterka

+0

我已經添加了代碼。 OpenCL:http://pastebin.com/cjGYSLQf CUDA:http://pastebin.com/VQMp3Hba –

+1

您的openCL代碼不包含內核源代碼... – talonmies

回答

1

你可以嘗試,並在代碼中插入明確定時器,而不是相信從工具的輸出。可能是該工具錯誤的情況。

+0

我已經完成了。我在CUDA和OpenCL中使用GPU計時器來測量內核的執行時間。它們與Nsight給我的結果完全一樣!所以這就是爲什麼我相信NSight! :) –