我試圖寫在OpenCL的直方圖內核來計算256倉R,G,和B的直方圖。我的內核是這樣的:OpenCL的圖像直方圖
const sampler_t mSampler = CLK_NORMALIZED_COORDS_FALSE |
CLK_ADDRESS_CLAMP|
CLK_FILTER_NEAREST;
__kernel void computeHistogram(read_only image2d_t input, __global int* rOutput,
__global int* gOutput, __global int* bOutput)
{
int2 coords = {get_global_id(0), get_global_id(1)};
float4 sample = read_imagef(input, mSampler, coords);
uchar rbin = floor(sample.x * 255.0f);
uchar gbin = floor(sample.y * 255.0f);
uchar bbin = floor(sample.z * 255.0f);
rOutput[rbin]++;
gOutput[gbin]++;
bOutput[bbin]++;
}
當我的2100 X 894的圖像上運行(1877400個像素)我往往只在看到或正在錄製的周圍1,870,000總價值時,我總結直方圖值對每個通道。每次都是不同的數字。我確實希望這樣做,因爲偶爾有兩個內核可能會從輸出數組中獲取相同的值並增加它,從而有效地取消一個增量操作(我假設?)。
的1,870,000輸出爲{1,1}工作組規模(這是什麼似乎得到默認設置,如果我沒有另行指定)。如果我強制像{10,6}這樣的較大的工作組大小,我會在直方圖中得到一個非常小的總和(與工作組大小的變化成正比)。這似乎很奇怪,我,但我猜會發生什麼情況是,所有的組中的工作項目的同時增加輸出數組值,所以它只是被計爲一次增量?
不管怎麼說,我讀過在OpenCL的沒有全局內存同步化,只有使用他們的__local內存本地工作組內同步化規範。由NVIDIA直方圖示例打破了直方圖工作量成一束特定大小的子問題,計算它們的局部直方圖,然後將結果合併到後一個單一的直方圖。對於任意大小的圖像來說,這看起來不會很好。我想我可以用虛擬值填充圖像數據...
作爲OpenCL的新手,我想我想知道是否有更直接的方法來做到這一點(因爲它看起來應該是一個相對簡單的方法GPGPU問題)。
謝謝!