开辟内存

OpenCL 内存对象就是一些 OpenCL 数据，这些数据一般在设备内存中，能够被拷入也能够被拷出。 OpenCL 内存对象包括 buffer 对象和 image 对象。
? Buffer 对象：连续的内存块 —-顺序存储，能够通过指针、行列式等直接访问。
? Image 对象：是 2 维或 3 维的内存对象，只能通过 read_image() 或 write_image() 来读取。 image 对象可以是可读或可写的，但不能同时既可读又可写。
该函数会在指定的 context 上创建一个 buffer 对象，image 对象相对比较复杂，留在后面再讲。 flags 参数指定 buffer对象的读写属性，host_ptr 可以是 NULL，如果不为 NULL，一般是一个有效的 host buffer 对象，这时，函数创建 OpenCL buffer 对象后，会把对应 host buffer 的内容拷贝到 OpenCL buffer 中。
在 Kernel 执行之前，host 中原始输入数据必须显式的传到 device 中，Kernel 执行完后，结果也要从 device 内存中传回到 host 内存中。我们主要通过函数 clEnqueue{Read/Write}Buffer/Image} 来实现这两种操作。从 host 到 device，我们用 clEnqueueWrite，从 device 到 host，我们用 clEnqueueRead。 clEnqueueWrite 命令包括初始化内存对象以及把host 数据传到 device 内存这两种操作。当然，像前面一段说的那样，也可以把 host buffer 指针直接用在 CreateBuffer 函数中来实现隐式的数据写操作。

/*Step 7: Initial input,output for the host and create memory objects for the kernel*/ const char* input = "GdkknVnqkc"; size_t strlength = strlen(input); cout << "input string:" << endl; cout << input << endl; char *output = (char*) malloc(strlength + 1); cl_mem inputBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY|CL_MEM_COPY_HOST_PTR, (strlength + 1) * sizeof(char),(void *) input, NULL); cl_mem outputBuffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY , (strlength + 1) * sizeof(char), NULL, NULL);

创建Kernel对象

在这里要特别说明，”“

/*Step 8: Create kernel object */ cl_kernel kernel = clCreateKernel(program,"vecadd", NULL);

设置Kernel参数

这里的参数设置就是传给kernel的参数，0，1，2就是顺序，sizeof就是类型，还有一个就是存在从机上的地址。

/*Step 9: Sets Kernel arguments.*/ cl_int clnum = BUFSIZE; status = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void*) &clbuf1); status = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void*) &clbuf2); clSetKernelArg(kernel, 2, sizeof(cl_mem), (void*) &buffer);

运行Kernel

//执行kernel,Range用1维，work itmes size 为BUFSIZE，没有设置group size，这时候，系统会使用默认的work group size。
size_t global_work_size = BUFSIZE;
status = clEnqueueNDRangeKernel( queue, kernel, 1,
NULL,&global_work_size, NULL, 0, NULL, &ev);

/*Step 10: Running the kernel.*/ size_t global_work_size = BUFSIZE; status = clEnqueueNDRangeKernel( queue, kernel, 1, NULL,&global_work_size, NULL, 0, NULL, &ev);

拷贝结果回主机

/*Step 11: Read the cout put back to host memory.*/ cl_float *ptr; cl_event mapevt; ptr = (cl_float *) clEnqueueMapBuffer( queue, buffer,CL_TRUE, CL_MAP_READ, 0, BUFSIZE * sizeof(cl_float), 0, NULL, NULL, NULL );

释放资源

/*Step 12: Clean the resources.*/ status = clReleaseKernel(kernel); //Release kernel. status = clReleaseProgram(program); //Release the program object. status = clReleaseMemObject(clbuf1); status = clReleaseMemObject(clbuf2); status = clReleaseMemObject(buffer); status = clReleaseCommandQueue(commandQueue); //Release Command queue. status = clReleaseContext(context); //Release context. if (buffer != NULL) { free(clbuf1); output = NULL; } if (devices != NULL) { free(devices); devices = NULL; }

完整代码

kernel

__kernel void vecadd(__global const float* A, __global const float* B, __global
float* C)
{
int id = get_global_id(0);
C[id] = A[id] + B[id];
}

c++