网站结构设计的内容,开创云网站建设支持,做网站建设销售途径,推广普通话黑板报原文http://www.olcf.ornl.gov/training_articles/opencl-vector-addition/ 本文仅仅是为了学习OpenCL而做的的相关翻译。 由于原文中的例子不能在我的环境中运行#xff0c;因此做了一些改动。 通过这个例子能很好地了解OpenCL的编程模型。 1. 简介 这个例子是表示了两个向量…原文http://www.olcf.ornl.gov/training_articles/opencl-vector-addition/ 本文仅仅是为了学习OpenCL而做的的相关翻译。 由于原文中的例子不能在我的环境中运行因此做了一些改动。 通过这个例子能很好地了解OpenCL的编程模型。 1. 简介 这个例子是表示了两个向量相加可以认为是OpenCL中的hello world。为了使程序更容易理解没有加入错误处理机制。 //vecAdd.c#include stdio.h
#include stdlib.h
#include math.h
#include CL/opencl.h// OpenCL kernel. Each work item takes care of one element of c
const char *kernelSource \n \
__kernel void vecAdd( __global float *a, \n \__global float *b, \n \__global float *c, \n \const unsigned int n) \n \
{ \n \//Get our global thread ID \n \int id get_global_id(0); \n \\n \//Make sure we do not go out of bounds \n \if (id n) \n \c[id] a[id] b[id]; \n \
} \n \\n ;int main( int argc, char* argv[] )
{// 向量长度int n 8;// 输入向量int *h_a;int *h_b;// 输出向量int *h_c;// 设备输入缓冲区cl_mem d_a;cl_mem d_b;// 设备输出缓冲区cl_mem d_c;cl_platform_id cpPlatform; // OpenCL 平台cl_device_id device_id; // device IDcl_context context; // contextcl_command_queue queue; // command queuecl_program program; // programcl_kernel kernel; // kernel//每个向量的字节数size_t bytes n*sizeof(int);//为每个向量分配内存h_a (int*)malloc(bytes);h_b (int*)malloc(bytes);h_c (int*)malloc(bytes);//初始化向量int i;for( i 0; i n; i ){h_a[i] i;h_b[i] i;}size_t globalSize, localSize;cl_int err;//每个工作组的工作节点数目localSize 2;//所有的工作节点globalSize (size_t)ceil(n/(float)localSize)*localSize;printf(%d\n,globalSize);//获得平台IDerr clGetPlatformIDs(1, cpPlatform, NULL);//获得设备ID与平台有关err clGetDeviceIDs(cpPlatform, CL_DEVICE_TYPE_CPU, 1, device_id, NULL);//根据设备ID得到上下文context clCreateContext(0, 1, device_id, NULL, NULL, err);//根据上下文在设备上创建命令队列queue clCreateCommandQueue(context, device_id, 0, err);//根据OpenCL源程序创建计算程序program clCreateProgramWithSource(context, 1,(const char **) kernelSource, NULL, err);//创建可执行程序clBuildProgram(program, 0, NULL, NULL, NULL, NULL);//在上面创建的程序中创建内核程序kernel clCreateKernel(program, vecAdd, err);//分配设备缓冲d_a clCreateBuffer(context, CL_MEM_READ_ONLY, bytes, NULL, NULL);d_b clCreateBuffer(context, CL_MEM_READ_ONLY, bytes, NULL, NULL);d_c clCreateBuffer(context, CL_MEM_WRITE_ONLY, bytes, NULL, NULL);// 将向量信息写入设备缓冲err clEnqueueWriteBuffer(queue, d_a, CL_TRUE, 0,bytes, h_a, 0, NULL, NULL);err | clEnqueueWriteBuffer(queue, d_b, CL_TRUE, 0,bytes, h_b, 0, NULL, NULL);// 设置计算内核的参数err clSetKernelArg(kernel, 0, sizeof(cl_mem), d_a);err | clSetKernelArg(kernel, 1, sizeof(cl_mem), d_b);err | clSetKernelArg(kernel, 2, sizeof(cl_mem), d_c);err | clSetKernelArg(kernel, 3, sizeof(int), n);// 在数据集的范围内执行内核Execute the kernel over the entire range of the data seterr clEnqueueNDRangeKernel(queue, kernel, 1, NULL, globalSize, localSize,0, NULL, NULL);// 在读出结果之前等待命令队列执行完毕Wait for the command queue to get serviced before reading back resultsclFinish(queue);// 从设备缓冲区读出结果Read the results from the deviceclEnqueueReadBuffer(queue, d_c, CL_TRUE, 0,bytes, h_c, 0, NULL, NULL );//输出读出的结果float sum 0;for(i0; in; i)printf(%d ,h_c[i]);// 释放资源clReleaseMemObject(d_a);clReleaseMemObject(d_b);clReleaseMemObject(d_c);clReleaseProgram(program);clReleaseKernel(kernel);clReleaseCommandQueue(queue);clReleaseContext(context);//释放内存free(h_a);free(h_b);free(h_c);system(pause);return 0;
}2. 基本解释 2.1 内核 Kernel是OpenCL代码的核心。全部kernel必须要作为一个C字符串读入最容易的方式就是将整个kernel用引号包起来行尾回车。在真正的程序中应该将kernel放在一个独立的文件中。 // OpenCL kernel. Each work item takes care of one element of c
const char *kernelSource \n \
__kernel void vecAdd( __global float *a, \n \__global float *b, \n \__global float *c, \n \const unsigned int n) \n \
{ \n \//Get our global thread ID \n \int id get_global_id(0); \n \\n \//Make sure we do not go out of bounds \n \if (id n) \n \c[id] a[id] b[id]; \n \
} \n \\n ;查看一下这个简单的内核由什么内容组成 __kernel void vecAdd( __global float *a, __global float *b, __global float *c, const unsigned int n) __kernel 指明这是一个OpenCL内核__global 说明指针指向的是全局的设备内存空间其它的就是C语言的函数的语法。kernel必须返回空类型。 int id get_global_id(0); 得到第0维全局工作节点的ID。 if (id n) c[id] a[id] b[id]; 工作组的数目必须是一个整数或者每个工作组的工作节点数目必须能被全部工作节点数目整除。由于共组的的大小被用来协调性能没有必要一定能被所有线程数目整除所以通常启用的线程比所需要的线程多一些并忽略掉多余的。在考察了问题域之后就能访问、操作设备内存了。2.2 内存 // 输入向量 int *h_a;int *h_b;// 输出向量 int *h_c;// 设备输入缓冲区 cl_mem d_a; cl_mem d_b;// 设备输出缓冲区 cl_mem d_c; CPU和GPU有不同的内存空间所以必须支持对内存分别引用一个集市主机数组指针另外一个集是设备内存的操作句柄。这儿我们用 h_和d_前缀来区分。 2.3 线程映射 //每个工作组的工作节点数目 localSize 2;//所有的工作节点 globalSize (size_t)ceil(n/(float)localSize)*localSize; 为了将问题映射到底层硬件必须指明局部的大小和全局的大小。局部大小定义了工作组中节点的数目子NVIDIA GPU上这相当于线程块内线程的数目。全局大小定义了所有启动的工作节点数目。localSize大小必须能被globalSize整除所以我们计算了一个最小的整数能覆盖问题域并且能被localSize整除。 2.4 环境配置 //绑定平台 err clGetPlatformIDs(1, cpPlatform, NULL); 每个硬件提供商都有不同的平台在用之前就应该给定这儿clGetPlatformIDs()会将cpPlatform赋予系统可用的平台。例如如果系统包含了AMD CPU和NVIDIA GPU且这两个平台都安装了合适的OpenCL驱动那这里平台都是可用的。注要使用不同的平台驱动必须安装相关的驱动在本例中我安装了AMD(ATI)的app SDK v2.5和Intel的intel_ocl_sdk_1.5_runtime_setup所以会有两个平台但是由于我的ATI的显卡GPU不能被app SDK v2.5支持所以的获得设备ID时没有用GPU设备而是用了CPU设备。如果这里配置不正确下面的可能就无法进行 //获得设备ID与平台有关err clGetDeviceIDs(cpPlatform, CL_DEVICE_TYPE_CPU, 1, device_id, NULL); 可以查询平台来得到它包含什么样的设备。在这个例子中用枚举值CL_DEVICE_TYPE_CPU来查询平台上的CPU设备。 //根据上下文在设备上创建命令队列 queue clCreateCommandQueue(context, device_id, 0, err); 在使用OpenCL设备之前必须要配置contextcontext被用来管理命令队列内存和内核的活动。一个context可以包含不止一个设备。 //根据上下文在设备上创建命令队列 queue clCreateCommandQueue(context, device_id, 0, err); 命令队列被用来将命令从主机放入指定的设备。内存的转移和内核的活动都能被放入命令队列在合适的时候在指定的设备上执行。 2.5 编译内核 //根据OpenCL源程序创建计算程序 program clCreateProgramWithSource(context, 1, (const char **) kernelSource, NULL, err);//创建可执行程序 clBuildProgram(program, 0, NULL, NULL, NULL, NULL);//在上面创建的程序中创建内核程序 kernel clCreateKernel(program, vecAdd, err); 为了保证对于大多数设备的可移植性默认运行内核的方式就是用即时Just-in-time编译我们必须为给定上下文的设备准备源码。首先创建程序这是一个内核程序的的集合然后根据程序来创建各自的内核程序。 2.6 准备数据 //分配设备缓冲 d_a clCreateBuffer(context, CL_MEM_READ_ONLY, bytes, NULL, NULL); d_b clCreateBuffer(context, CL_MEM_READ_ONLY, bytes, NULL, NULL); d_c clCreateBuffer(context, CL_MEM_WRITE_ONLY, bytes, NULL, NULL);// 将向量信息写入设备缓冲 err clEnqueueWriteBuffer(queue, d_a, CL_TRUE, 0, bytes, h_a, 0, NULL, NULL); err | clEnqueueWriteBuffer(queue, d_b, CL_TRUE, 0, bytes, h_b, 0, NULL, NULL);// 设置计算内核的参数 err clSetKernelArg(kernel, 0, sizeof(cl_mem), d_a); err | clSetKernelArg(kernel, 1, sizeof(cl_mem), d_b); err | clSetKernelArg(kernel, 2, sizeof(cl_mem), d_c); err | clSetKernelArg(kernel, 3, sizeof(int), n); 在启动内核之前必须在设备和主机之间创建缓冲区将主机数据绑定到新创建的设备缓冲区上最后设置内核参数。 2.7 启动内核 // 在数据集的范围内执行内核 err clEnqueueNDRangeKernel(queue, kernel, 1, NULL, globalSize, localSize,0, NULL, NULL); 一旦内存驻留在设备上之后内核就能排队启动了。 2.8 取回结果 // 在读出结果之前等待命令队列执行完毕 clFinish(queue);// 从设备缓冲区读出结果 clEnqueueReadBuffer(queue, d_c, CL_TRUE, 0, bytes, h_c, 0, NULL, NULL ); 可以进行阻断直到所有的命令队列执行完毕然后将设备上的结果取回到主机。 3. 运行环境 3.1 OpenCL: AMD app sdk v2.5 intel_ocl_sdk_1.5_runtime 3.2 Visual Studio 2010 express转载于:https://www.cnblogs.com/wangshide/archive/2011/11/04/2235204.html
