使用OpenCL编程时，kernel写成一个单独的文件或者将文件内容保存在一个string中。可以使用clBuildProgram对kernel进行编译链接(compiles & links)，如果失败，可以使用clGetProgramBuildInfo获取OpenCL编译器对kernel的编译信息。
1.clBuildProgram
cl_int clBuildProgram (
cl_program program, //program
cl_uint num_devices, //the number of device
const cl_device_id *device_list, //devices id
const char *options, //the option of compiler
void (CL_CALLBACK *pfn_notify)(cl_program program, void *user_data), //the callback function
void *user_data) //the data of callback function
)
2.clGetProgramBuildInfo
cl_int clGetProgramBuildInfo (
cl_program program, //program
cl_device_id device, //the id of device
cl_program_build_info param_name,
size_t param_value_size,
void *param_value,
size_t *param_value_size_ret
)
3.代码实例(获取编译器对kernel的编译信息)
3.1 kernel(build_info_kernel.cl)
1 __kernel void good(__global float *a, 2 __global float *b, 3 __global float *c) { 4 5 *c = *a + *b; 6 } 7 8 __kernel void good(__global float *a, 9 __global float *b, 10 __global float *c) { 11 __local int var=3; 12 int size=get_local_sze(0); 13 *c = *a + *b; 14 }
3.2 tool.h
1 #ifndef TOOLH 2 #define TOOLH 3 #include <CL/cl.h> 4 #include <string.h> 5 #include <stdio.h> 6 #include <stdlib.h> 7 #include <iostream> 8 #include <string> 9 #include <fstream> 10 using namespace std; 11 12 /** convert the kernel file into a string */ 13 int convertToString(const char *filename, std::string& s); 14 15 /**Getting platforms and choose an available one.*/ 16 int getPlatform(cl_platform_id &platform); 17 18 /**Step 2:Query the platform and choose the first GPU device if has one.*/ 19 cl_device_id *getCl_device_id(cl_platform_id &platform); 20 21 /**获取编译program出错时，编译器的出错信息*/ 22 int getProgramBuildInfo(cl_program program,cl_device_id device); 23 #endif
tool.cpp
1 #include <CL/cl.h> 2 #include <string.h> 3 #include <stdio.h> 4 #include <stdlib.h> 5 #include <iostream> 6 #include <string> 7 #include <fstream> 8 #include "tool.h" 9 using namespace std; 10 11 /** convert the kernel file into a string */ 12 int convertToString(const char *filename, std::string& s) 13 { 14 size_t size; 15 char* str; 16 std::fstream f(filename, (std::fstream::in | std::fstream::binary)); 17 18 if(f.is_open()) 19 { 20 size_t fileSize; 21 f.seekg(0, std::fstream::end); 22 size = fileSize = (size_t)f.tellg(); 23 f.seekg(0, std::fstream::beg); 24 str = new char[size+1]; 25 if(!str) 26 { 27 f.close(); 28 return 0; 29 } 30 31 f.read(str, fileSize); 32 f.close(); 33 str[size] = ' '; 34 s = str; 35 delete[] str; 36 return 0; 37 } 38 cout<<"Error: failed to open file :"<<filename<<endl; 39 return -1; 40 } 41 42 /**Getting platforms and choose an available one.*/ 43 int getPlatform(cl_platform_id &platform) 44 { 45 platform = NULL;//the chosen platform 46 47 cl_uint numPlatforms;//the NO. of platforms 48 cl_int status = clGetPlatformIDs(0, NULL, &numPlatforms); 49 if (status != CL_SUCCESS) 50 { 51 cout<<"Error: Getting platforms!"<<endl; 52 return -1; 53 } 54 55 /**For clarity, choose the first available platform. */ 56 if(numPlatforms > 0) 57 { 58 cl_platform_id* platforms = 59 (cl_platform_id* )malloc(numPlatforms* sizeof(cl_platform_id)); 60 status = clGetPlatformIDs(numPlatforms, platforms, NULL); 61 platform = platforms[0]; 62 free(platforms); 63 } 64 else 65 return -1; 66 } 67 68 /**Step 2:Query the platform and choose the GPU device*/ 69 cl_device_id *getCl_device_id(cl_platform_id &platform) 70 { 71 cl_uint numDevices = 0; 72 cl_device_id *devices=NULL; 73 cl_int status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &numDevices); 74 if (numDevices > 0) //GPU available. 75 { 76 devices = (cl_device_id*)malloc(numDevices * sizeof(cl_device_id)); 77 status = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, numDevices, devices, NULL); 78 } 79 return devices; 80 } 81 82 /**获取编译program出错时，编译器的出错信息*/ 83 int getProgramBuildInfo(cl_program program,cl_device_id device) 84 { 85 size_t log_size; 86 char *program_log; 87 /* Find size of log and print to std output */ 88 clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 89 0, NULL, &log_size); 90 program_log = (char*) malloc(log_size+1); 91 program_log[log_size] = ' '; 92 clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 93 log_size+1, program_log, NULL); 94 printf("%s ", program_log); 95 free(program_log); 96 return 0; 97 }
3.3 buildInfo.cpp
1 #include "tool.h" 2 #include <string.h> 3 #include <stdio.h> 4 #include <stdlib.h> 5 #include <iostream> 6 #include <string> 7 #include <fstream> 8 using namespace std; 9 10 void CL_CALLBACK checkData(cl_program platform, void* data){ 11 printf("%s ",(char*)data); 12 } 13 14 int main(int argc, char* argv[]) 15 { 16 cl_int status; 17 /** Getting platforms and choose an available one(first).*/ 18 cl_platform_id platform; 19 getPlatform(platform); 20 21 /**Query the platform and choose the GPU device.*/ 22 cl_device_id *devices=getCl_device_id(platform); 23 24 /**Create context use the frist device.*/ 25 cl_context context = clCreateContext(NULL,1, devices,NULL,NULL,NULL); 26 27 /**Create program object */ 28 const char *filename = "build_info_kernel.cl"; 29 string sourceStr; 30 status = convertToString(filename, sourceStr); 31 const char *source = sourceStr.c_str(); 32 size_t sourceSize[] = {strlen(source)}; 33 cl_program program = clCreateProgramWithSource(context, 1, &source, sourceSize, NULL); 34 35 /**Build program. */ 36 //status=clBuildProgram(program, 1,devices,NULL,checkData,"sdf"); 37 status=clBuildProgram(program, 1,devices,NULL,NULL,NULL); 38 if(status < 0) //get the build info 39 getProgramBuildInfo(program ,devices[0]); 40 else 41 printf("Build Success "); 42 43 status = clReleaseProgram(program); //Release the program object. 44 status = clReleaseContext(context);//Release context. 45 free(devices); 46 47 getchar(); 48 return 0; 49 }
对kernel的编译结果：
