啓動大內核大小時出現未知錯誤

Question

當我的數組大小超過591（591）時啓動簡單內核時遇到問題。大小爲591x591時，數組返回時沒有任何錯誤，但是儘快我以每個16x16線程的38x38塊的網格維度啓動內核，內核無法啓動並返回「未知錯誤」。

下面的代碼是我打電話內核和在我的代碼調用內核：

#include <cuda.h> 
#include <cuda_runtime.h> 
#include <cuda_device_runtime_api.h> 

using namespace std; 

#define BLOCKSIZE 16 
#define gpuErrchk(ans) { gpuAssert((ans), __FILE__,__LINE__);} 

inline void gpuAssert(cudaError_t code, char *file, int line, bool abort = true) 
{ 
if (code != cudaSuccess) 
{ 
    fprintf(stderr, "GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line); 
    if(abort) exit(code); 
} 
} 



__global__ void IdentityMatrixKernel(float* identity, int size) 
{ 
int index_x = blockIdx.x * blockDim.x + threadIdx.x; 
int index_y = blockIdx.y * blockDim.y + threadIdx.y; 

// map the two 2D indices to a single linear, 1D index 
int grid_width = gridDim.x * blockDim.x; 
int index = index_y * grid_width + index_x; 

// map the two 2D block indices to a single linear, 1D block index 
//int result = blockIdx.y * gridDim.x + blockIdx.x; 


if (index % (size+1)) 
{ 
    identity[index] = 0; 
} 
else 
{ 
    identity[index] = 1; 
} 


void foo(float *aArray, int size) 
{ 
float* d_I; 
int size2 = size*size*sizeof(float); 

gpuErrchk(cudaMalloc(&d_I,size2)); 

dim3 block_size; 
block_size.x = BLOCKSIZE; 
block_size.y = BLOCKSIZE; 

dim3 grid_size; 
grid_size.x = size1/ block_size.x + 1; 
grid_size.y = size1/ block_size.y + 1; 

IdentityMatrixKernel<<<grid_size,block_size>>>(d_I,size); 
gpuErrchk(cudaPeekAtLastError()); 

gpuErrchk(cudaMemcpy(aArray,d_I,size2,cudaMemcpyDeviceToHost)); 

cudaFree(d_I); 
} 

int main() 
{ 
int size = 591; 
float *aArray = (float*)malloc(size*size*sizeof(float)); 

foo(aArray,size); 


return 0; 
} 

對於size = 591沒有錯誤顯示出來，輸出尺寸591x591的單位矩陣，但對於任何較大尺寸它吐出向控制檯輸出「未知錯誤」。

Answer 1

一個問題似乎是，你正在啓動線程的一個網格，比你的實際矩陣較大：

grid_size.x = size1/ block_size.x + 1; 
grid_size.y = size1/ block_size.y + 1; 

但你沒有檢查出任何的越界訪問您的內核。你需要添加一個線程檢查如：

if ((index_x >= size)||(index_y >= size)) return; 

接近你的內核的開始。但這還不夠。另一個問題是，你的index計算不正確：

int index = index_y * grid_width + index_x; 

在它的表面上，這似乎是正確的，但是因爲你的線程陣列比你的數據陣列（可能）較大，這可以提供不正確的分度。既然你傳遞size內核無論如何，改變它的東西是這樣的：

int index = index_y * size + index_x; 

，你應該能夠消除出界外訪問。

Answer 2

我擴展了Robert Crovella的答案。

如果你定義了一個大數字的block_size。{x，y}（在你的情況16），那麼你將不能處理更小的數組，例如4×4。你可以做的是定義一個小塊大小：

/* create thread blocks */ 
dim3 block_size; 
block_size.x = 4; 
block_size.y = 4; 

/* create n x n block grids */ 
dim3 grid_size; 
grid_size.x = size1/block_size.x; 
grid_size.y = size1/block_size.y; 

/* in case of partial sizes make grid_size 1 x 1 */ 
if (size1 % block_size.x) 
    grid_size.x = 1, grid_size.y = 1;