OpenCl_CPU加速矩阵运算

本博文用的是intel的opencl架构，下载链接https://software.intel.com/en-us/intel-opencl/download，默认安装即可

注意：安装完毕后opencl的sdk在路径C:\Program Files (x86)\Intel\OpenCL SDK\6.3下

第一步：检验计算机硬件设备

安装完毕检验硬件设备，查看平台数量，代码如下：

[cpp] view plain copy

1. #include <iostream>  
2. #include <malloc.h>  
3. #include <CL/cl.h>//包含CL的头文件  
4.   
5. using namespace std;  
6.   
7. //根据参数，判断设备类别。是CPU、GPU、ACCELERATOR或其他设备  
8. const char* GetDeviceType(cl_device_type it)  
9. {  
10.     if (it == CL_DEVICE_TYPE_CPU)  
11.         return "CPU";  
12.     else if (it == CL_DEVICE_TYPE_GPU)  
13.         return "GPU";  
14.     else if (it == CL_DEVICE_TYPE_ACCELERATOR)  
15.         return "ACCELERATOR";  
16.     else  
17.         return "DEFAULT";  
18.   
19. }  
20.   
21. int main()  
22. {  
23.     char dname[512];  
24.     cl_device_id devices[20];  
25.     cl_platform_id* platform_id = NULL;  
26.     cl_uint num_devices;  
27.     cl_device_type int_type;  
28.     cl_ulong long_entries;  
29.     cl_uint num_platform;  
30.     cl_int err;  
31.   
32.     //查询系统上可用的计算平台，可以理解为初始化  
33.     err = clGetPlatformIDs(0, NULL, &num_platform);  
34.   
35.     if (err != CL_SUCCESS)  
36.     {  
37.         cout << "clGetPlatformIDs error" << endl;  
38.         return 0;  
39.     }  
40.   
41.     cout << "PlatForm num:" << num_platform << endl;  
42.   
43.     int st = 0;  
44.   
45.     platform_id = new cl_platform_id[num_platform];  
46.   
47.     err = clGetPlatformIDs(num_platform, platform_id, NULL);  
48.   
49.     if (err != CL_SUCCESS)  
50.     {  
51.         cout << "clGetPlatformIDs error" << endl;  
52.         return 0;  
53.     }  
54.   
55.     for (st = 0; st<num_platform; st++)  
56.     {  
57.         cout << "----------------------------------" << endl;  
58.         cout << "Platform " << st + 1 << endl;  
59.   
60.         //获取可用计算平台的名称  
61.         clGetPlatformInfo(platform_id[st], CL_PLATFORM_NAME, 512, dname, NULL);  
62.         cout << "CL_PLATFORM_NAME:" << dname << endl;  
63.   
64.         //获取可用计算平台的版本号,即OpenCL的版本号  
65.         clGetPlatformInfo(platform_id[st], CL_PLATFORM_VENDOR, 512, dname, NULL);  
66.         cout << "CL_PLATFORM_VERSION:" << dname << endl;  
67.   
68.         //获取可用计算平台的设备数目  
69.         clGetDeviceIDs(platform_id[st], CL_DEVICE_TYPE_ALL, 20, devices, &num_devices);  
70.         cout << "Device num:" << num_devices << endl;  
71.   
72.         unsigned int n = 0;  
73.   
74.         //循环两次，检测两个设备的属性  
75.         for (n = 0; n<num_devices; n++)  
76.         {  
77.             cout << endl << "Device " << n + 1 << endl;  
78.             //获取设备名称  
79.             clGetDeviceInfo(devices[n], CL_DEVICE_NAME, 512, dname, NULL);  
80.             cout << "Device :" << dname << endl;  
81.   
82.             //获取设备类别  
83.             clGetDeviceInfo(devices[n], CL_DEVICE_TYPE, sizeof(cl_device_type),  
84.                 &int_type, NULL);  
85.             cout << "Device Type:" << GetDeviceType(int_type) << endl;  
86.   
87.             //获取设备版本号  
88.             clGetDeviceInfo(devices[n], CL_DRIVER_VERSION, 512, dname, NULL);  
89.             cout << "Device version:" << dname << endl;  
90.   
91.             //获取设备全局内存大小  
92.             clGetDeviceInfo(devices[n], CL_DEVICE_GLOBAL_MEM_SIZE,  
93.                 sizeof(cl_ulong), &long_entries, NULL);  
94.             cout << "Device global mem(MB):" <<  
95.                 long_entries / 1024 / 1024 << endl;  
96.   
97.             //获取设备CACHE内存大小  
98.             clGetDeviceInfo(devices[n], CL_DEVICE_GLOBAL_MEM_CACHE_SIZE,  
99.                 sizeof(cl_ulong), &long_entries, NULL);  
100.             cout << "Device global mem cache(KB):" <<  
101.                 long_entries / 1024 << endl;  
102.   
103.             //获取本地内存大小  
104.             clGetDeviceInfo(devices[n], CL_DEVICE_LOCAL_MEM_SIZE,  
105.                 sizeof(cl_ulong), &long_entries, NULL);  
106.             cout << "Device Locale mem(KB) :" << long_entries / 1024 << endl;  
107.   
108.             //获取设备频率  
109.             clGetDeviceInfo(devices[n], CL_DEVICE_MAX_CLOCK_FREQUENCY,  
110.                 sizeof(cl_ulong), &long_entries, NULL);  
111.             cout << "Device Max clock(MHz) :" << long_entries << endl;  
112.   
113.             //获取最大工作组数  
114.             clGetDeviceInfo(devices[n], CL_DEVICE_MAX_WORK_GROUP_SIZE,  
115.                 sizeof(cl_ulong), &long_entries, NULL);  
116.             cout << "Device Max Group size :" << long_entries << endl;  
117.   
118.             //获取最大计算核心数  
119.             clGetDeviceInfo(devices[n], CL_DEVICE_MAX_COMPUTE_UNITS,  
120.                 sizeof(cl_ulong), &long_entries, NULL);  
121.             cout << "Device Max parallel cores:" << long_entries << endl;  
122.   
123.         }  
124.     }  
125.   
126.     return 0;  
127. }  

如果发现cpu的opencl版本为实验版本2.1，需要runtime配置文件，安装即可。

配置文件链接：http://pan.baidu.com/s/1geNnMy3 密码：7n8o

再次检验硬件设备，查看平台数量，可以发现cpu的opencl版本正确。

第二步：矩阵运算加速

代码如下：

核函数vecadd.cl文件如下：

[cpp] view plain copy

1. __kernel void vecAdd(__global int* A,  
2.         __global int* B,  
3.         __global int* C)  
4. {  
5.     //获取当前工作项所在位置（线程索引号）  
6.     //就是向量每一维的位置  
7.     int idx = get_global_id(0);  
8.     C[idx] = A[idx] + B[idx];  
9. }  

主函数main.cpp文件如下：

[cpp] view plain copy

1. #include <iostream>  
2. #include <stdio.h>  
3. #include <string.h>  
4. #include <string>  
5. #include <vector>  
6. #include <CL/cl.h>//包含CL的头文件  
7.   
8. //OpenCl 2.1  
9. //Solve problem: error C4996: 'clCreateCommandQueue': 被声明为已否决  
10. #pragma warning( disable : 4996 )  
11.   
12. using namespace std;  
13.   
14. //100维向量  
15. #define elements 100  
16.   
17. //从外部文件获取cl内核代码  
18. bool GetFileData(const char* fname, string& str)  
19. {  
20.     FILE* fp = fopen(fname, "r");  
21.     if (fp == NULL)  
22.     {  
23.         printf("no found filen");  
24.         return false;  
25.     }  
26.   
27.     int n = 0;  
28.     while (feof(fp) == 0)  
29.     {  
30.         str += fgetc(fp);  
31.     }  
32.   
33.     return true;  
34. }  
35.   
36. int main()  
37. {  
38.     //先读外部CL核心代码，如果失败则退出。  
39.     //代码存buf_code里面  
40.     string code_file;  
41.   
42.     if (false == GetFileData("vecadd.cl", code_file))  
43.     {  
44.         return 0;  
45.     }  
46.   
47.     char* buf_code = new char[code_file.size()];  
48.     strcpy(buf_code, code_file.c_str());  
49.     buf_code[code_file.size() - 1] = NULL;  
50.   
51.     //声明CL所需变量。  
52.     cl_device_id device;  
53.     cl_platform_id *platform_id = NULL;  
54.     cl_context context;  
55.     cl_command_queue cmdQueue;  
56.     cl_mem bufferA, bufferB, bufferC;  
57.     cl_program program;  
58.     cl_kernel kernel = NULL;  
59.   
60.     //我们使用的是一维向量  
61.     //设定向量大小（维数）  
62.     size_t globalWorkSize[1];  
63.     globalWorkSize[0] = elements;  
64.   
65.     cl_int err;  
66.   
67.     //定义输入变量和输出变量，并设定初值  
68.     int* buf_A = new int[elements];  
69.     int* buf_B = new int[elements];  
70.     int* buf_C = new int[elements];  
71.   
72.     size_t datasize = sizeof(int) * elements;  
73.   
74.     for (int i = 0; i < elements; i++)  
75.     {  
76.         buf_A[i] = (float)i;  
77.         buf_B[i] = (float)i + 1.0;  
78.     }  
79.   
80.     //step 1:初始化OpenCL  
81.     cl_uint num_platform;  
82.     err = clGetPlatformIDs(0, NULL, &num_platform);  
83.   
84.     platform_id = new cl_platform_id[num_platform];  
85.     err = clGetPlatformIDs(num_platform, platform_id, NULL);  
86.   
87.     if (err != CL_SUCCESS)  
88.     {  
89.         cout << "clGetPlatformIDs error" << endl;  
90.         return 0;  
91.     }  
92.   
93.     //博主计算机三个plantform，platform_id[2]为CPU，根据情况来改  
94.     clGetDeviceIDs(platform_id[2], CL_DEVICE_TYPE_CPU, 1, &device, NULL);  
95.   
96.     //step 2:创建上下文  
97.     context = clCreateContext(NULL, 1, &device, NULL, NULL, NULL);  
98.   
99.     //step 3:创建命令队列  
100.     cmdQueue = clCreateCommandQueue(context, device, 0, NULL);  
101.   
102.     //step 4:创建数据缓冲区  
103.     bufferA = clCreateBuffer(context, CL_MEM_READ_ONLY, datasize, NULL, NULL);  
104.     bufferB = clCreateBuffer(context, CL_MEM_READ_ONLY, datasize, NULL, NULL);  
105.     bufferC = clCreateBuffer(context, CL_MEM_WRITE_ONLY, datasize, NULL, NULL);  
106.   
107.     //step 5:将数据上传到缓冲区  
108.     clEnqueueWriteBuffer(cmdQueue, bufferA, CL_FALSE, 0, datasize, buf_A, 0, NULL, NULL);  
109.     clEnqueueWriteBuffer(cmdQueue, bufferB, CL_FALSE, 0, datasize, buf_B, 0, NULL, NULL);  
110.   
111.     //step 6:加载编译代码,创建内核调用函数  
112.     program = clCreateProgramWithSource(context, 1, (const char**)&buf_code, NULL, NULL);  
113.     clBuildProgram(program, 1, &device, NULL, NULL, NULL);  
114.     kernel = clCreateKernel(program, "vecAdd", NULL);  
115.   
116.     //step 7:设置参数，执行内核  
117.     clSetKernelArg(kernel, 0, sizeof(cl_mem), &bufferA);  
118.     clSetKernelArg(kernel, 1, sizeof(cl_mem), &bufferB);  
119.     clSetKernelArg(kernel, 2, sizeof(cl_mem), &bufferC);  
120.   
121.     clEnqueueNDRangeKernel(cmdQueue, kernel, 1, NULL, globalWorkSize, NULL, 0, NULL, NULL);  
122.   
123.     //step 8:取回计算结果  
124.     clEnqueueReadBuffer(cmdQueue, bufferC, CL_TRUE, 0, datasize, buf_C, 0, NULL, NULL);  
125.   
126.     //输出验证结果  
127.     cout << buf_A[0] << "+" << buf_B[0] << "=" << buf_C[0] << endl;  
128.     cout << buf_A[elements - 1] << "+" << buf_B[elements - 1] << "=" << buf_C[elements - 1] << endl;  
129.   
130.     //释放所有调用和内存  
131.     clReleaseKernel(kernel);  
132.     clReleaseProgram(program);  
133.     clReleaseCommandQueue(cmdQueue);  
134.     clReleaseMemObject(bufferA);  
135.     clReleaseMemObject(bufferB);  
136.     clReleaseMemObject(bufferC);  
137.     clReleaseContext(context);  
138.   
139.     delete[]platform_id;  
140.     delete[]buf_A;  
141.     delete[]buf_B;  
142.     delete[]buf_C;  
143.     delete[]buf_code;  
144.     system("pause");  
145.     return 0;  
146. }  

测试结果显示如下：