OpenCL中kernel的迴圈呼叫

kernel的迴圈呼叫主要是涉及緩衝區的建立和主機端命令同步；很多例子中會在建立快取物件時對快取物件做初始化，例如：

cl_mem memobject1 = clcreatebuffer(context,cl_mem_read_only |

cl_mem_copy_host_ptr ,
sizeof(float) * mixsize,&input1,&error);

其一 在本次實驗中，我們需要將快取物件的建立和資料傳輸分開執行，例如：

cl_mem memobject1 = clcreatebuffer(context,cl_mem_read_only ,

sizeof(float) * mixsize,null,&error);
error = clenqueuewritebuffer(queue, memobject1, cl_false, 0, 
mixsize * sizeof(float), a_in, 0, null, null);

host.c

#include

#include
#include
#pragma warning( disable : 4996 )
#define mixsize 8192
int main() 
error = clgetdeviceids(platforms, cl_device_type_gpu, 1, &devices, null);
if (error != 0) 
//建立上下文
context = clcreatecontext(null,1,&devices,null,null,&error);
if (error != 0) 
//建立程式
program_handle = fopen("kernel.cl","rb");
if (program_handle == null) 
fseek(program_handle,0,seek_end);
program_size = ftell(program_handle);
rewind(program_handle);
program_buffer = (char *)malloc(program_size+1);
program_buffer[program_size] = '\0';
error=fread(program_buffer,sizeof(char),program_size,program_handle);
if (error == 0) 
fclose(program_handle);
program = clcreateprogramwithsource(context,1,(const
char **)&program_buffer,&program_size,&error);
if (error < 0) 
//編譯程式
error = clbuildprogram(program,1,&devices,null,null,null);
if (error < 0) 
//建立命令佇列
queue = clcreatecommandqueue(context, devices, cl_queue_profiling_enable, &error);
if (error < 0) 
//建立核心
kernel = clcreatekernel(program,kernel_name,&error);
if (kernel==null) 
//建立快取物件
cl_mem memobject1 = clcreatebuffer(context,cl_mem_read_only ,
sizeof(float) * mixsize,null,&error);
if (error < 0) 
cl_mem memobject2 = clcreatebuffer(context, cl_mem_read_only , 
sizeof(float) * mixsize, null, &error);
if (error < 0) 
cl_mem memobject3 = clcreatebuffer(context, cl_mem_read_write , sizeof(float) * mixsize, null, &error);
if (error < 0) 
//設定核心引數
error = clsetkernelarg(kernel,0,sizeof(cl_mem),&memobject1);
error|= clsetkernelarg(kernel, 1, sizeof(cl_mem), &memobject2);
error |= clsetkernelarg(kernel, 2, sizeof(cl_mem), &memobject3);
if (error != cl_success) 
//該實驗是kernel迴圈呼叫；將建立緩衝物件和資料寫入緩衝物件分開執行
//kernel有三個輸入引數
//實現的功能：
//a和b更新10次，result用於將上一次運算的結果與自身資料加和，並負責輸出最終結果；
//初始化引數
float result[mixsize];
float a_in[mixsize];
float b_in[mixsize];
float c_in[mixsize];
cl_int status = 0;
cl_event evt1;
cl_event evt2;
cl_event evt3;
for (int i = 0; i < mixsize; i++) 
//配置工作項
size_t maxworkgroupsize = 0;
clgetdeviceinfo(devices, cl_device_max_work_group_size,
sizeof(maxworkgroupsize), &maxworkgroupsize, null);
size_t globalworksize = mixsize;
size_t localworksize = maxworkgroupsize;
for (int j = 0; j < 10; j++) 
error = clenqueuewritebuffer(queue, memobject2, cl_false, 0, mixsize * sizeof(float), b_in, 1, &evt1, &evt2);
if (error != cl_success) 
for (int i = 0; i < mixsize; i++) 
//執行核心
error = clenqueuendrangekernel(queue, kernel, 1, null, &globalworksize, &localworksize, 1, &evt2, &evt3);
if (error != cl_success) 
//資料初始化
for (int i = 0; i < mixsize; i++) 
//同步
clwaitforevents(1, &evt3);
clreleaseevent(evt1);
clreleaseevent(evt2);
clreleaseevent(evt3);
}//讀取執行結果
error = clenqueuereadbuffer(queue,memobject3,cl_true,0,mixsize*sizeof(float),result,0,null,null);
if (error != cl_success) 
//顯示結果
for (int i = 0; i < mixsize; i++) 
}printf("successed!\n");
clreleaseprogram(program);
clreleasecontext(context);
clreleasecommandqueue(queue);
clreleasedevice(devices);
clreleasekernel(kernel);
getchar();
return
0;}

kernel.cl

//資料加法

__kernel void createbuffer(__global const
float *a_in,
__global const
float *b_in,
__global float *result) 

迴圈呼叫，所以memobject3的屬性需要設定為讀寫；

OpenCL在kernel檔案中加入第三方標頭檔案等

kernel編寫過程中，想在.cl檔案中加入第三方的庫檔案標頭檔案，除了在clbuildprogram 中修改options選項以外，還需要在kernel檔案頭包含相關檔案。例子如下 cpp檔案中 int main int argc,char argv const char options i f ...

opencl中裝置記憶體

local float x 4.0 這樣會報錯，解決辦法是 local float x x 4.0 private 可以限定可以用於限定核心引數以及所有非內涵函式的引數和變數。slsetkernelarg kernel,0,16 sizeof float null 對應的核心函式 kernel vo...

kernel中的per cpu變數

per cpu 變數的引入有效的解決了smp系統中處理器對鎖得競爭，每個cpu只需訪問自己的本地變數。本文闡述了per cpu變數在2.6核心上的實現和相關操作。在系統編譯階段我們就手工的定義了乙份所有的per cpu變數，這些變數的定義是通過巨集define per cpu實現的 11 defin...