Как сделать ядро ​​гистограммы CUDA?

Question

Я пишу ядро ​​CUDA для гистограммы на картинке, но я понятия не имел, как вернуть массив из ядра, и массив изменится, когда другой поток прочитает его. Любое возможное решение для этого?

__global__ void Hist(
    TColor *dst, //input image 
    int imageW, 
    int imageH, 
int*data 
){ 
    const int ix = blockDim.x * blockIdx.x + threadIdx.x; 
    const int iy = blockDim.y * blockIdx.y + threadIdx.y; 

if(ix < imageW && iy < imageH) 
{ 
    int pixel = get_red(dst[imageW * (iy) + (ix)]); 
        //this assign specific RED value of image to pixel 

    data[pixel] ++; // ?? problem statement ... 
} 
} 

@para d_dst: входное изображение TColor равно float4.

@para данные: массив для размера гистограммы [255]

extern "C" void 
cuda_Hist(TColor *d_dst, int imageW, int imageH,int* data) 
{ 
    dim3 threads(BLOCKDIM_X, BLOCKDIM_Y); 
    dim3 grid(iDivUp(imageW, BLOCKDIM_X), iDivUp(imageH, BLOCKDIM_Y)); 
    Hist<<<grid, threads>>>(d_dst, imageW, imageH, data); 
} 

Answer 1

Вы посмотрели образец SDK? Образец «гистограммы» доступен в CUDA SDK (в настоящее время версия 3.0 на NVIDIA developer site, версия 3.1 beta доступна для registered developers).

В документации по образцу хорошо объясняется, как обрабатывать ваше суммирование, используя атомизацию глобальной памяти на графическом процессоре или путем сбора результатов для каждого блока отдельно, а затем делая отдельное сокращение (либо на хосте, либо на графическом процессоре).

Answer 2

гистограммирования не является особенно эффективным, когда осуществляется с CUDA (или с GPGPU в целом) - как правило, вам необходимо сгенерировать множество частичных гистограмм в и затем суммировать их. Возможно, вы захотите рассмотреть эту задачу на CPU.

Answer 3

Вам придется либо использовать атомную функцию, чтобы заблокировать другой поток от использования его той же памяти, либо использовать частичную гистограмму. В любом случае это не так эффективно, если входное изображение очень велико.