如何實現opencl內核管道

Question

我正在使用OpenCl開展我的項目。爲了提高我的算法的性能，是否可以管道一個內核？如果一個內核由多個步驟組成，讓我們說A，B，C，我希望A在完成它的部分並將它傳遞給B時接受新數據。我可以在它們之間創建通道，但是我不知道如何執行它詳細。

我可以在.cl文件中寫入A，B，C（3個內核）嗎？但如何入列NDRange？ 我正在使用Altera SDK進行FPGA HPC開發。 謝謝。

Answer 1

流水線可以通過使用幾個與通道連接的內核來實現。隨着所有內核同時運行，數據從一個轉移到另一個：

Pipeline example from Intel FPGA OpenCL SDK Programming Guide

該管道非常簡單的例子是：

channel int foo_bar_channel; 
channel float bar_baz_channel; 

__kernel void foo(__global int* in) { 
    for (int i = 0; i < 1024; ++i) { 
    int value = in[i]; 
    value = clamp(value, 0, 255);     // do some work 
    write_channel_altera(foo_bar_channel, value); // send data to the next kernel 
    } 
} 

__kernel void bar() { 
    for (int i = 0; i < 1024; ++i) { 
    int value = read_channel_altera(foo_bar_channel); // take data from foo 
    float fvalue = (float) value; 
    write_channel_altera(bar_baz_channel, value); // send data to the next kernel 
    } 
} 

__kernel void baz(__global int* out) { 
    for (int i = 0; i < 1024; ++i) {n 
    float value = read_channel_altera(bar_baz_channel); 
    float s = sin(value); 
    out[i] = s;         // write result in the end 
    } 
} 

你可以寫在一個單一的.CL所有內核文件或使用不同的文件，然後＃將它們包括到主.cl文件中。

我們希望我們所有的內核可以同時運行，因此它們可以接受彼此的數據。由於只有有序命令隊列支持，我們必須使用不同的隊列，每個內核：

cl_queue foo_queue = clCreateCommandQueue(...); 
cl_queue bar_queue = clCreateCommandQueue(...); 
cl_queue baz_queue = clCreateCommandQueue(...); 

clEnqueueTask(foo_queue, foo_kernel); 
clEnqueueTask(bar_queue, bar_kernel); 
clEnqueueTask(baz_queue, baz_kernel); 

clFinish(baz_queue); // last kernel in our pipeline 

不同的OpenCL編程GPU，我們依靠數據流水線，所以NDRange內核不會給我們帶來任何好處。使用單個工作項內核而不是NDRange內核，所以我們使用clEnqueueTask函數對它們進行排隊。額外的內核屬性（reqd_work_group_size）可用於標記單個工作項內核，爲編譯器提供一些優化空間。

檢查有關渠道和內核屬性（具體而言，第1.6.4節針對OpenCL渠道推廣實施英特爾FPGA SDK）的更多信息，英特爾FPGA SDK針對OpenCL編程指南：

https://www.altera.com/en_US/pdfs/literature/hb/opencl-sdk/aocl_programming_guide.pdf