OpenCL标准(OpenCL standard):
OpenCL定义了含有C++的C的API接口,可以用于绑定第三方语言:Java,python,.NET。OpenCL C严格执行了C99的标准!
OpenCL说明(OpenCL specification):
OpenCL包括四个部分:
1、平台模型:有主机和设备(设备上就是跑OpenCL)。在设备上跑的OpenCL代码就叫做kernels。
2、执行模型:定义了OpenCL在主机上的环境配置以及kernel在设备上如何执行。包括了诸如在主机上创建一个OpenCL项目,提供主机-设备交互,并且定义了在设备上执行kernel时的并发模型。
3、内存模型:定义了kernel使用的抽象内存层次,即使在真的内存架构下。内存模型与GPU内存层次很接近。
4、编程模型:定义了映射到物理硬件的并发模型。
OpenCL的执行模型和kernel:
一个普通的向量加法:
// Perform an element-wise addition of A and B and store in C.
// There are N elements per array.
void vecadd(int *C, int* A, int *B, int N) {
for(int i ¼ 0; i < N; i++) {
C[i] = A[i] + B[i];
}
}
对于多核处理器的做法是:(strip mining)
// Perform and element-wise addition of A and B and store in C.
// There are N elements per array and NP CPU cores.
void vecadd(int *C, int* A, int *B, int N, int NP, int tid) {
int ept ¼ N/NP; // elements per thread
for(int i = tid*ept; i < (tid+1)*ept; i++) {
C[i] = A[i] + B[i];
}
}
OpenCL的做法:(work-item,n维排序)
// Perform an element-wise addition of A and B and store in C
// N work-items will be created to execute this kernel.
__kernel
void vecadd(__global int *C, __global int* A, __global int *B) {
int tid ¼ get_global_id(0); // OpenCL intrinsic function
C[tid] ¼ A[tid] + B[tid];
}
size_t indexSpaceSize[3]={1024, 1, 1}; //size_t就是定义了该维的大小
Work item swith in a work group have a special relationship with one another:They can perform barrier operations to synchronize and they have access to a shared memory address space.size_t workGroupSize[3]={64, 1, 1};如果每个数组的大小是1024,那么就会有1024/64 = 16个group
OpenCL的平台和设备:
设备是计算单元,可以随便变化的
OpenCL的执行环境:
cl_int
clGetPlatformIDs(cl_uint num_entries,
cl_platform_id *platforms,
cl_uint *num_platforms)
cl_int
clGetPlatformIDs(cl_uint num_entries,
cl_platform_id *platforms,
cl_uint *num_platforms)
OpenCL的内存模型:
OpenCL的kernel的书写:
AMD OPENCL的SDK: http://developer.amd.com/tools-and-sdks/
入门教程:(非常好用)http://www.kimicat.com/opencl-1/zai-windows-xia-shi-yong-opencl
接下来就说一下OpenCL的程序的执行步骤步骤:
翻译出来就是:
(1)Discover
and initialize the platforms调用两次clGetPlatformIDs函数,第一次获取可用的平台数量,第二次获取一个可用的平台。
(2)Discover
and initialize the devices调用两次clGetDeviceIDs函数,第一次获取可用的设备数量,第二次获取一个可用的设备。
(3)Create
a context(调用clCreateContext函数)上下文context可能会管理多个设备device。
(4)Create
a command queue(调用clCreateCommandQueue函数)一个设备device对应一个command
queue。
上下文conetxt将命令发送到设备对应的command
queue,
设备就可以执行命令队列里的命令。
(5)Create
device buffers(调用clCreateBuffer函数)
Buffer中保存的是数据对象,就是设备执行程序需要的数
据保存在其中。Buffer由上下文conetxt创建,这样上下文管理的多个设备就会共享Buffer中的数据。
(6)Write
host data to device buffers:(调用clEnqueueWriteBuffer函数)
(7)Create
and compile the program:创建程序对象,程序对象就代表你的程序源文件或者二进制代码数据。
(8)Create
the kernel(调用clCreateKernel函数):根据你的程序对象,生成kernel对象,表示设备程序的
入口。
(9)Set
the kernel arguments(调用clSetKernelArg函数)
(10)Configure
the work-item structure(设置worksize):配置work-item的组织形式(维数,group组成等)
(11)Enqueue
the kernel for execution:(调用clEnqueueNDRangeKernel函数)将kernel对象,以及
work-item参数放入命令队列中进行
执行。
(12)Read
the output buffer back to the host:(调用clEnqueueReadBuffer函数)
(13)Release
OpenCL resources:(至此结束整个运行过程)