OpenCL编程基本流程及完整示例

Nicole_Terasic

        1. 选择OpenCL平台并创建一个上下文

        平台（Platform）是指主机和OpenCL管理框架下的若干个设备构成的可以运行OpenCL程序的完整硬件系统，这个是跑OpenCL程序的基础，所以第一步要选择一个可用的OpenCL品台。一台机器上可以有不止一个这样的平台，一个平台也可以有不止一个GPU。

        主要涉及的函数： clGetPlatformIDs（），用于获取可用的平台；
        clCreateContextFromType（）， 创建一个OpenCL运行时山下文环境；

        2. 选择设备并创建命令队列

        选择平台并创建好OpenCL上下文环境之后，要做的事选择运行时用到的设备，还要创建一个命令队列，命令队列里定义了设备要完成的操作，以及各个操作的运行次序。
        主要涉及的函数：clCreateCommandQueue（），用于创建一个指定设备上的上下文环境，第二个参数定义了选择的设备。

        3. 创建和构建程序对象

        程序对象用来存储与上下文相关联的设备的已编译可执行代码，同时也完成内核源代码的加载编译工作。
        主要涉及的函数：clCreateProgramWithSource（）， 这个函数会创建一个程序对象，在创建的同时，把已经转化成字符串形式的内核源代码加载到该程序对象中。
        clBuildProgram（）用于编译指定程序对象中的内核源代码，编译成功之后，再把编译代码存储在程序对象中。

        4. 创建内核和内存对象

        要执行程序对象中的已编译成功的内核运算，需要在内存中创建内核并分配内核函数的参数，在GPU上定义内存对象并分配存储空间。
        主要涉及的函数：clCreateKernel（）， 创建内核；
        clCreateBuffer（），分配内存对象的存储空间，这些对象可以由内核函数直接访问。

        5. 设置内核数据并执行内核

        创建内核和内存对象之后，接下来要设置核函数的数据，并将要执行的内核排队。
        主要涉及的函数：clEnqueueNDRangeKernel（），用于设置内核函数的所有参与运算的数据。利用命令队列对要在设备上执行的内核排队。需要注意的是，执行内核排队之后并不意味着这个内核一定会立即执行，只是排队到了执行队列中。


        6. 读取执行结果并释放OpenCL资源

        内核执行完成之后，需要把数据从GPU拷贝到CPU中，供主机进一步处理，所有者写工作完成之后需要释放所有的OpenCL资源。
        主要涉及的函数：clEnqueueReadBuffer（），读取设备内存数据到主机内存；
        clReleaseXXX（），释放OpenCL资源。

        以下程序包含了以上所有6个步骤，功能很简单，实现两个数组求和。

主程序：

1. #include <iostream>
   
2. #include <fstream>
   
3. #include <sstream>
   
4. #include <CL/cl.h>
   
5. 
   
6. const int ARRAY_SIZE = 1000;
   
7. 
   
8. //一、 选择OpenCL平台并创建一个上下文
   
9. cl_context CreateContext()
   
10. {
    
11.         cl_int errNum;
    
12.         cl_uint numPlatforms;
    
13.         cl_platform_id firstPlatformId;
    
14.         cl_context context = NULL;
    
15. 
    
16.         //选择可用的平台中的第一个
    
17.         errNum = clGetPlatformIDs(1, &firstPlatformId, &numPlatforms);
    
18.         if (errNum != CL_SUCCESS || numPlatforms <= 0)
    
19.         {
    
20.                 std::cerr << "Failed to find any OpenCL platforms." << std::endl;
    
21.                 return NULL;
    
22.         }
    
23. 
    
24.         //创建一个OpenCL上下文环境
    
25.         cl_context_properties contextProperties[] =
    
26.         {
    
27.                 CL_CONTEXT_PLATFORM,
    
28.                 (cl_context_properties)firstPlatformId,
    
29.                 0
    
30.         };
    
31.         context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
    
32.                 NULL, NULL, &errNum);
    
33. 
    
34.         return context;
    
35. }
    
36. 
    
37. 
    
38. //二、 创建设备并创建命令队列
    
39. cl_command_queue CreateCommandQueue(cl_context context, cl_device_id *device)
    
40. {
    
41.         cl_int errNum;
    
42.         cl_device_id *devices;
    
43.         cl_command_queue commandQueue = NULL;
    
44.         size_t deviceBufferSize = -1;
    
45. 
    
46.         // 获取设备缓冲区大小
    
47.         errNum = clGetContextInfo(context, CL_CONTEXT_DEVICES, 0, NULL, &deviceBufferSize);
    
48. 
    
49.         if (deviceBufferSize <= 0)
    
50.         {
    
51.                 std::cerr << "No devices available.";
    
52.                 return NULL;
    
53.         }
    
54. 
    
55.         // 为设备分配缓存空间
    
56.         devices = new cl_device_id[deviceBufferSize / sizeof(cl_device_id)];
    
57.         errNum = clGetContextInfo(context, CL_CONTEXT_DEVICES, deviceBufferSize, devices, NULL);
    
58. 
    
59.         //选取可用设备中的第一个
    
60.         commandQueue = clCreateCommandQueue(context, devices[0], 0, NULL);
    
61. 
    
62.         *device = devices[0];
    
63.         delete[] devices;
    
64.         return commandQueue;
    
65. }
    
66. 
    
67. 
    
68. // 三、创建和构建程序对象
    
69. cl_program CreateProgram(cl_context context, cl_device_id device, const char* fileName)
    
70. {
    
71.         cl_int errNum;
    
72.         cl_program program;
    
73. 
    
74.         std::ifstream kernelFile(fileName, std::ios::in);
    
75.         if (!kernelFile.is_open())
    
76.         {
    
77.                 std::cerr << "Failed to open file for reading: " << fileName << std::endl;
    
78.                 return NULL;
    
79.         }
    
80. 
    
81.         std::ostringstream oss;
    
82.         oss << kernelFile.rdbuf();
    
83. 
    
84.         std::string srcStdStr = oss.str();
    
85.         const char *srcStr = srcStdStr.c_str();
    
86.         program = clCreateProgramWithSource(context, 1,
    
87.                 (const char**)&srcStr,
    
88.                 NULL, NULL);
    
89. 
    
90.         errNum = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
    
91. 
    
92.         return program;
    
93. }
    
94. 
    
95. //创建和构建程序对象
    
96. bool CreateMemObjects(cl_context context, cl_mem memObjects[3],
    
97.         float *a, float *b)
    
98. {
    
99.         memObjects[0] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
    
100.                 sizeof(float) * ARRAY_SIZE, a, NULL);
     
101.         memObjects[1] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
     
102.                 sizeof(float) * ARRAY_SIZE, b, NULL);
     
103.         memObjects[2] = clCreateBuffer(context, CL_MEM_READ_WRITE,
     
104.                 sizeof(float) * ARRAY_SIZE, NULL, NULL);
     
105.         return true;
     
106. }
     
107. 
     
108. 
     
109. // 释放OpenCL资源
     
110. void Cleanup(cl_context context, cl_command_queue commandQueue,
     
111.         cl_program program, cl_kernel kernel, cl_mem memObjects[3])
     
112. {
     
113.         for (int i = 0; i < 3; i++)
     
114.         {
     
115.                 if (memObjects[i] != 0)
     
116.                         clReleaseMemObject(memObjects[i]);
     
117.         }
     
118.         if (commandQueue != 0)
     
119.                 clReleaseCommandQueue(commandQueue);
     
120. 
     
121.         if (kernel != 0)
     
122.                 clReleaseKernel(kernel);
     
123. 
     
124.         if (program != 0)
     
125.                 clReleaseProgram(program);
     
126. 
     
127.         if (context != 0)
     
128.                 clReleaseContext(context);
     
129. }
     
130. 
     
131. int main(int argc, char** argv)
     
132. {
     
133.         cl_context context = 0;
     
134.         cl_command_queue commandQueue = 0;
     
135.         cl_program program = 0;
     
136.         cl_device_id device = 0;
     
137.         cl_kernel kernel = 0;
     
138.         cl_mem memObjects[3] = { 0, 0, 0 };
     
139.         cl_int errNum;
     
140. 
     
141.         // 一、选择OpenCL平台并创建一个上下文
     
142.         context = CreateContext();
     
143. 
     
144.         // 二、 创建设备并创建命令队列
     
145.         commandQueue = CreateCommandQueue(context, &device);
     
146. 
     
147.         //创建和构建程序对象
     
148.         program = CreateProgram(context, device, "HelloWorld.cl");
     
149. 
     
150.         // 四、 创建OpenCL内核并分配内存空间
     
151.         kernel = clCreateKernel(program, "hello_kernel", NULL);
     
152. 
     
153.         //创建要处理的数据
     
154.         float result[ARRAY_SIZE];
     
155.         float a[ARRAY_SIZE];
     
156.         float b[ARRAY_SIZE];
     
157.         for (int i = 0; i < ARRAY_SIZE; i++)
     
158.         {
     
159.                 a[i] = (float)i;
     
160.                 b[i] = (float)(ARRAY_SIZE - i);
     
161.         }
     
162. 
     
163.         //创建内存对象
     
164.         if (!CreateMemObjects(context, memObjects, a, b))
     
165.         {
     
166.                 Cleanup(context, commandQueue, program, kernel, memObjects);
     
167.                 return 1;
     
168.         }
     
169. 
     
170.         // 五、 设置内核数据并执行内核
     
171.         errNum = clSetKernelArg(kernel, 0, sizeof(cl_mem), &memObjects[0]);
     
172.         errNum |= clSetKernelArg(kernel, 1, sizeof(cl_mem), &memObjects[1]);
     
173.         errNum |= clSetKernelArg(kernel, 2, sizeof(cl_mem), &memObjects[2]);
     
174. 
     
175.         size_t globalWorkSize[1] = { ARRAY_SIZE };
     
176.         size_t localWorkSize[1] = { 1 };
     
177. 
     
178.         errNum = clEnqueueNDRangeKernel(commandQueue, kernel, 1, NULL,
     
179.                 globalWorkSize, localWorkSize,
     
180.                 0, NULL, NULL);
     
181. 
     
182.         // 六、 读取执行结果并释放OpenCL资源
     
183.         errNum = clEnqueueReadBuffer(commandQueue, memObjects[2], CL_TRUE,
     
184.                 0, ARRAY_SIZE * sizeof(float), result,
     
185.                 0, NULL, NULL);
     
186. 
     
187.         for (int i = 0; i < ARRAY_SIZE; i++)
     
188.         {
     
189.                 std::cout << result[i] << " ";
     
190.         }
     
191.         std::cout << std::endl;
     
192.         std::cout << "Executed program succesfully." << std::endl;
     
193.         getchar();
     
194.         Cleanup(context, commandQueue, program, kernel, memObjects);
     
195. 
     
196.         return 0;
     
197. }

复制代码

核函数文件“HelloWorld.cl”：

• __kernel void hello_kernel(__global const float *a,
•         __global const float *b,
•         __global float *result)
• {
•         int gid = get_global_id(0);
•         result[gid] = a[gid] + b[gid];
• }
  

执行结果：

图1

转自：https://www.cnblogs.com/mtcnn/p/9411859.html
作者：未雨愁眸