pycuda共享內存錯誤「pycuda._driver.LogicError：cuLaunchKernel失敗：無效值」

Question

我有一個奇怪的問題，我不能確定哪個產地：

我有一個工作的內核，並獲得一些特殊的矩陣向量乘法，這我想加快速度。基本上，大矩陣（10^6×10^6）由少量小矩陣構成。所以我想把這些數據放在共享內存中。然而，當我嘗試添加共享內存，我只得到了錯誤：

pycuda._driver.LogicError：cuLaunchKernel失敗：無效值

所以我的工作核心是：

#define FIELD_SIZE {field} 
#define BLOCK_SIZE {block} 

__global__ void MatrixMulKernel(double *gpu_matrix, double *gpu_b, double *gpu_y) 
{ 
    int tx = ... + threadIdx.x; 

    if(tx < FIELD_SIZE*FIELD_SIZE*BLOCK_SIZE) 
    { ... multiplication ... } 
} 

如果我嘗試添加它看起來像

#define FIELD_SIZE {field} 
#define BLOCK_SIZE {block} 

__global__ void MatrixMulKernel(double *gpu_matrix_ptr, double *gpu_b, double *gpu_y) 
{ 
    __shared__ double gpu_matrix[BLOCK_SIZE*BLOCK_SIZE*13]; 

    int tx = ... + threadIdx.x; 
    if(tx < BLOCK_SIZE*BLOCK_SIZE*13) { gpu_matrix[tx] = gpu_matrix_ptr[tx]; } 
    __syncthreads(); 

    if(tx < FIELD_SIZE*FIELD_SIZE*BLOCK_SIZE) 
    { ... multiplication ... } 
} 

這是我唯一改變的部分共享內存部分，所以基本上它是gpu_matrix [TX] = gpu_matrix_ptr [TX]聲明，不是招它呢？但我不明白應該如何。我基本上試圖從pycuda示例中複製平鋪的矩陣乘法示例。 http://wiki.tiker.net/PyCuda/Examples/MatrixmulTiled

的調用是：

self.kernel.prepare([np.intp, np.intp, np.intp]) 
self.kernel.prepared_call(grid_shape, 
       block_shape, 
       self.matrix_gpu.gpudata, 
       b_gpu.gpudata, 
       y_gpu.gpudata) 

其中matrix_gpu，b_gpu和y_gpu是pycuda.gpuarray實例。

希望你能澄清一些我的困惑......

Answer 1

根據你的描述，共享MEM您分配太大。

__shared__ double gpu_matrix[BLOCK_SIZE*BLOCK_SIZE*13]; 

shared mem是cuda gpu的硬件資源之一。總大小約爲48KBytes，你不能增加。

CUDA實際上在以下目錄中提供了一個工具來幫助您計算可以使用的硬件資源。

$CUDA_ROOT/tools/CUDA_Occupancy_Calculator.xls 

在另一方面，通過墊-VEC-MUL狀內核需要共享MEM的尺寸應該能夠從O（BLOCK_SIZE^2）降低到O（BLOCK_SIZE）。你可能想在實現你自己的代碼之前閱讀一些成功的mat-vec-mul內核的代碼，如MAGMA。