OpenCL平台API : 平台API定义了宿主机程序发现OpenCL设备所用的函数以及这些函数的功能,另外还定义了为OpenCL应用创建上下文的函数.这里推荐一下我的参考书籍：《OpenCL编程指南》和《OpenCL异构计算》，前者风格像一个说明文档，后者带你实践；推荐后者因为代码更简洁一些，本来OpenCL初始化都有点复杂了，而前者还能搞的更复杂，苦笑……，不过《指南》的优点是对API解释的比较详细。再送一个在线的OpenCL API查询库，建议收藏：OpenCL 2.1 Reference Pages让了让萌新们更快上手，我直接以实例开始，争取读完就能写下自己第一个OpenCL程序，至于API的学习参考书籍或者上一条为方便查阅，本系列用到的几个API将在篇二给出说明，我先写好的篇三ヾ(o◕∀◕)ﾉヾ。

一、初始化的平台相关变量

这里主要有平台、上下文、设备、命令队列、程序、内核等，这几项是通用的；其他如缓冲区等看需创建哈。文件开始先声明几个需要主要的变量：

cl_int errNum; cl_platform_id platform; // 0. 使用简明： cl_device_id device; // 1. 在平台选定设备 cl_context context; // 2.1 在指定设备创建上下文 cl_command_queue cmdqueue; // 创建一个命令队列关联上下文中指定设备 cl_program program; // 2.2 关联上下文创建程序对象,并对列表设备构建 cl_kernel kernel; // 3. 由程序对象创建内核 printf("\n\n OpenCL_Initing .....");

1.1、查询平台 clGetPlatformIDs( )

首先呢，我们申请一个平台变量的指针类型，在获取平台个数以后，再为其指定空间。对clGetPlatformIDs调用两次，第一次获取获取平台数量，第二次获取平台列表。有些老式设备在第一次调用clGetPlatformIDs可能花费时间较久，比如我的，建议选择以高性能显卡启动VS，情况会稍微缓解一些。

cl_uint numPlatforms =1; cl_platform_id *platformIds; errNum = clGetPlatformIDs( 1, NULL, &numPlatforms); platformIds =(cl_platform_id*)malloc(sizeof(cl_platform_id)*numPlatforms); errNum = clGetPlatformIDs( numPlatforms, platformIds, NULL); assert(errNum==CL_SUCCESS);

1.2、 从平台查询设备 clGetDeviceIDs( )

一般情况我们直接选定第一个平台，查询设备即可，入门可以选这样干着，日后熟悉OpenCL了可以实验深入一点的东西。这个方式和查询平台相同，第一次调用查询设备数，创建存储空间后第二次获取设备列表。这里需要指定设备类型。

platform = platformIds[0]; cl_uint numDevices =1; cl_device_id *deviceIds; errNum = clGetDeviceIDs( platform, CL_DEVICE_TYPE_GPU, 0, NULL, &numDevices); deviceIds =(cl_device_id*)malloc(sizeof(cl_device_id)*numDevices); errNum = clGetDeviceIDs( platform, CL_DEVICE_TYPE_GPU, numDevices, deviceIds, NULL); assert(errNum==CL_SUCCESS);

这之后一般建议打印设备信息，可以确认一下是不是指定了你想选择的设备,，下面三条一般足够：

//打印设备信息 char deviceinfo[100]; for(cl_uint i=0; i<numDevices; i++){ errNum =clGetDeviceInfo( deviceIds[i], CL_DEVICE_NAME, sizeof(deviceinfo), deviceinfo, 0 ); printf("\nDevice:%-2d CL_DEVICE_NAME: %s",i,deviceinfo); errNum =clGetDeviceInfo( deviceIds[i], CL_DEVICE_VENDOR, sizeof(deviceinfo), deviceinfo, 0 ); printf("\n %-2d CL_DEVICE_VENDOR: %s",i,deviceinfo); errNum =clGetDeviceInfo( deviceIds[i], CL_DEVICE_VERSION, sizeof(deviceinfo), deviceinfo, 0 ); printf("\n %-2d CL_DEVICE_VERSION: %s\n",i,deviceinfo); }

1.3、创建上下文 clCreateContext( )

上下文是所有OpenCL应用的核心。上下文为关联的设备、内存对象（例如，缓冲和图像）以及命令队列（在上下文和各设备之间提供一个接口）提供一个容器。 必须由一个平台的设备创建各个上下文，然后利用上下文可以做到：

创建一个或多个命令队列创建程序，使它在一个或者多个关联设备上运行从这些程序创建一个内核。在宿主机或设备上分配内存缓冲区图像将内核（设置适当的参数）提交到命令队列来执行，等 // 在平台选定设备 创建上下文 cl_context_properties props[3] = { CL_CONTEXT_PLATFORM,(cl_context_properties)platform, 0 }; context = clCreateContext( props, numDevices, deviceIds, NULL, NULL, &errNum); assert(errNum==CL_SUCCESS);

1.4、在上下文中针对设备创建命令队列 clCreateCommandQueue( )

在OpenCL中命令队列也是一个重要的概念，其实本篇文章中用到的都是相当重要的基础概念，因为以后程序运行中要执行的操作都要提交到命令队列。

device = deviceIds[0]; cmdqueue = clCreateCommandQueue( context, device, 0, &errNum); assert(errNum==CL_SUCCESS);

1.5、在上下文中创建程序对象

这里主要就是载入你的内核文件（一般是.cl结尾），只要这里构建成功，接下来就可以从program创建各种内核。KernelFileName就是你的文件名，如不在程序相同目录时，需要指定文件路径的。我的文件地址是KernelFileName =”./../MyKernel/ThisIsKernelFile.cl”；“./”表示当前工作目录，“../”表示上一级目录。从这一步开始（第5步）请同学们根据自己的程序改写代码。运行中可以看到srcStr就是内核文件。

// 载入程序文件 std::ifstream kernelFile(KernelFileName, std::ios::in); std::ostringstream oss; oss<<kernelFile.rdbuf(); kernelFile.close(); // 读取文件字符内容 std::string srcStdStr = oss.str(); const char *srcStr = srcStdStr.c_str(); program = clCreateProgramWithSource( context, 1, &srcStr, NULL, &errNum); // 程序对象在选定设备构建 errNum = clBuildProgram( program, 0, NULL, NULL, NULL, NULL); assert(errNum==CL_SUCCESS);

1.6、创建内核 clCreateKernel( )

这个API就比较简单好用了，参数设置也只有三个。这里需要保证你的程序文件包含以kernel_name命名的内核。

char *kernel_name ="My_Kernel_One" ; kernel =clCreateKernel( program, kernel_name, &errNum);

比如我的.cl文件部分代码：

// FLIE: ./../MyKernel/ThisIsKernelFile.cl __kernel void My_Kernel_One( __global uint * const input, const int inputWidth ) { const int x = get_global_id(0); const int y = get_global_id(1); output[y * get_global_size(0) + x] = (y * get_global_size(0) + x); }

二、一些额外的说明

2.1、在内核文件中调用函数

这个是可以的，因为我已经使用了；不过需要将函数写在内核文件（.cl）中。

2.2、在内核文件使用结构体

这个好像是可以的，因为函数调用都实现了；应该是仍然需要将结构体在.cl文件定义，这个仅仅在主机端定义是不行的，感觉,.cl文件是一个相对独立，不受主机端影响，只受主机端调用。

2.3、在内核中使用printf输出数据

这个也是可以的，因为我也使用了；不过需要在.cl文件开头，添加引用声明：

#pragma OPENCL EXTENSION cl_amd_printf : enable

2.4、一个在线调试内核文件的方法

刚开始学习OpenCL如果还没接触到调试工具，这里可以教你很好的办法调试.cl文件；用下面的可循环代码创建程序对象。一点建议：以下部分代码最好之printf英文，避免不必要的乱码。

while(1) { std::ifstream kernelFile(KernelFileName, std::ios::in); std::ostringstream oss; oss << kernelFile.rdbuf(); kernelFile.close(); std::string srcStdStr = oss.str(); const char *srcStr = srcStdStr.c_str(); program = clCreateProgramWithSource( context, 1, &srcStr, NULL, &errNum); // 为程序对象在选定设备构建 errNum = clBuildProgram( program, 0, NULL, NULL, NULL, NULL); // 构建成功则退出 if(errNum==CL_SUCCESS) break; // 构建失败则输出构建日志,方便调试 system("cls"); printf("Error: Failed to build program executable!\n"); size_t logsize; clGetProgramBuildInfo( program, device, CL_PROGRAM_BUILD_LOG, 0, NULL, &logsize); char* logbuf =(char*)calloc(logsize+1,sizeof(char)); clGetProgramBuildInfo( program, device, CL_PROGRAM_BUILD_LOG, logsize, logbuf, NULL); // 输出构建日志 printf("%s\n\n", logbuf); //exit(1); free(logbuf); printf("Please updata .cl file before Y,or N to exit: "); char ch; scanf(" %c",&ch); if(ch=='n' || ch=='N' || ch=='0') break; }

2.5、编写OpenCL建 议将平台初始化部分给独立出来

比如编写一对CL_init.h 和CL_init.cpp，将其中的变量和函数声明称外部变量，在其他文件引用即可，简化了main文件，使得主程序文件更加简洁，方便专注于算法，而且以后创建新的OpenCL程序也会很方便。

我的CL_init.cpp和CL_init.cpp，仅供参考。

2.6、想想还有啥好说的，待续……