opencl:clEnqueueNDRangeKernel執行報錯CL_OUT_OF_RESOURCES的一種情況

我的電腦上之前的顯示卡比較老并不支援opencl，是以我之前開發時opencl代碼其實都是在CPU上跑的，現在所有的代碼都調試通過了，決定裝塊新顯示卡用于程式的性能測試。

今天顯示卡到了，裝上之後運作程式，

clEnqueueNDRangeKernel

在執行下面的kernel時報錯：

CL_OUT_OF_RESOURCES

。

__kernel void prefix_sum_col_and_transpose( __constant SRC_TYPE *src, __global DST_TYPE * dst, uint width,uint height, uint src_width_step, uint dst_width_step){
    .........// 代碼實作部分略過
}
           

百撕不得其姐啊。。。。這代碼在CPU上跑很正常，邏輯沒問題呀。

最後發現隻是kernel 指針參數的位址修飾符使用不當造成的。

上面這段代碼，是用于圖像積分圖計算的，對給定的原圖(src)資料計算積分圖，輸出到目标指針(dst)指向的全局記憶體中。因為src資料不允許被修改是以我想當然的把src指定為

__constant

。

而這裡用

__constant

修飾是不對的。

__constant

和

__global

都是全局記憶體，

__constant

修飾的位址指向的是常量，不能被修改，但它們之間的差別卻并不僅于此。

一個opencl裝置的常量空間是有限制的，通過clGetDeviceInfo擷取

CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE

可以知道一個opencl裝置的最大常量緩沖區的尺寸，在我的顯示卡上，這個值是65536，簡單通過指令行運作AMD APP SDK的

clinfo

就可以得到這個值，如下圖：

因為圖像的尺寸很容易就超過64kb，是以

clEnqueueNDRangeKernel

在執行kernel時無法将它放到opencl裝置的constant buffer中，是以就會報錯

CL_OUT_OF_RESOURCES

。

是以應該将

src

的位址修飾符從

__constant

改為

__global

,如果要禁止修改

src

指針的資料，前面用c語言标準的

const

關鍵字修飾這個指針就可以了，是以這個kernel函數正确的定義應該是這樣：

__kernel void prefix_sum_col_and_transpose( const __global SRC_TYPE *src, __global DST_TYPE * dst, uint width,uint height, uint src_width_step, uint dst_width_step){
    .........// 代碼實作部分略過
}