OpenCL 第一个程序

我们使用矢量加操作作为第一个OpenCL程序,进入OpenCL编程讲解。

内核程序vecAdd.cl内容:

 _kernel void vector_add(global const float *a,
                             global const float *b, global float *result)
 {
     int gid = get_global_id(0);
     result[gid] = a[gid] + b[gid];
 }

主机端代码:

#include 〈stdio.h>
#include 〈stdlib.h>
#ifdef _APPLE_
#include 〈OpenCL/cl.h>
#else
#include 〈CL/cl.h>
#endif
const int ARRAY_SIZE = 1000;
char *ReadKernelSourceFile(const char *filename, size_t *length)
{
    FILE *file = NULL;
    size_t sourceLength;
    char *sourceString;
    int ret;
    file = fopen(filename, "rb");
    if(file == NULL)
 }
 {
    printf("%s at %d :Can't open %s\n", _FILE_, _LINE_ - 2,
            filename);
    return NULL;
 }
 fseek(file, 0, SEEK_END);
 sourceLength = ftell(file);
 fseek(file, 0, SEEK_SET);
 sourceString = (char *)malloc(sourceLength + 1)
                  sourceString[0] = '\0';
ret = fread(sourceString, sourceLength, 1, file);
 if(ret == 0)
 {
    printf("%s at %d : Can't read source %s\n", _FILE_,
            _LINE_ - 2, filename);
    return NULL;
 }
 fclose(file);
 if(length != 0)
 {
    *length = sourceLength;
 }
 sourceString[sourceLength] = '\0';
 return sourceString;
}
/*
1.创建平台
2.创建设备
3.根据设备创建上下文
*/
cl_context CreateContext(cl_device_id *device)
{
 cl_int errNum;
 cl_uint numPlatforms;
 cl_platform_id firstPlatformId;
 cl_context context = NULL;
 errNum = clGetPlatformIDs(1, &firstPlatformId, &numPlatforms);
 if (errNum != CL_SUCCESS || numPlatforms 〈= 0)
 {
    printf( "Failed to find any OpenCL platforms." );
    return NULL;
 }
 errNum = clGetDeviceIDs(firstPlatformId, CL_DEVICE_TYPE_GPU, 1,
                            device, NULL);
 if (errNum != CL_SUCCESS)
 {
    printf( "There is no GPU, trying CPU..." );
    errNum = clGetDeviceIDs(firstPlatformId,
                                  CL_DEVICE_TYPE_CPU, 1, device, NULL);
 }
 if (errNum != CL_SUCCESS)
 {
    printf( "There is NO GPU or CPU" );
    return NULL;
 }
context = clCreateContext(NULL, 1, *device, NULL, NULL, & errNum);
 if (errNum != CL_SUCCESS)
 {
    printf( " create context error\n" );
    return NULL;
 }
 return context;
}
 /*
 @在上下文可用的第一个设备中创建命令队列
 */
 cl_command_queue CreateCommandQueue(cl_context context,
                                          cl_device_id device)
{
    cl_int errNum;
    cl_command_queue commandQueue = NULL;
    commandQueue = clCreateCommandQueue(context, device, 0, NULL);
    if (commandQueue == NULL)
    {
        printf("Failed to create commandQueue for device 0");
        return NULL;
    }
    return commandQueue;
 }
 /*
 @读取内核源码创建OpenCL程序
 */
 cl_program CreateProgram(cl_context context,
                              cl_device_id device,
                              const char *fileName)
 {
    cl_int errNum;
    cl_program program;
    size_t program_length;
    char *const source = ReadKernelSourceFile (fileName,
                                                        &program_length);
    program = clCreateProgramWithSource(context, 1,
                                              (const char **)&source,
                                                NULL, NULL);
    if (program == NULL)
    {
        printf("Failed to create CL program from source." );
        return NULL;
    }
 errNum = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
 if (errNum != CL_SUCCESS)
 {
    char buildLog[16384];
    clGetProgramBuildInfo(program, device,
                              CL_PROGRAM_BUILD_LOG,
                              sizeof(buildLog),
                              buildLog, NULL);
    printf("Error in kernel:%s ", buildLog);
    clReleaseProgram(program);
    return NULL;
 }
 return program;
}
/*
@创建内存对象
*/
bool CreateMemObjects(cl_context context, cl_mem memObjects[3],
                    float * a, float * b)
{
 memObjects[0] = clCreateBuffer(context,
                        CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
                        sizeof(float) * ARRAY_SIZE, a, NULL);
 memObjects[1] = clCreateBuffer(context,
                        CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
                        sizeof(float) * ARRAY_SIZE, b, NULL);
 memObjects[2] = clCreateBuffer(context, CL_MEM_READ_WRITE,
                                    sizeof(float) * ARRAY_SIZE, NULL, NULL);
 if (memObjects[0] == NULL || memObjects[1] == NULL
    || memObjects[2] == NULL)
 {
    printf("Error creating memory objects." );
    return false;
   }
 return true;
}
/*
@清除OpenCL资源
*/
void Cleanup(cl_context context, cl_command_queue commandQueue,
          cl_program program, cl_kernel kernel,
          cl_mem memObjects[3])
{
 for (int i = 0; i 〈 3; i++)
 {
    if (memObjects[i] != 0)
        clReleaseMemObject(memObjects[i]);
 }
 if (commandQueue != 0)
    clReleaseCommandQueue(commandQueue);
 if (kernel != 0)
    clReleaseKernel(kernel);
 if (program != 0)
    clReleaseProgram(program);
 if (context != 0)
    clReleaseContext(context);
}
int main(int argc, char **argv)
{
 cl_context context = 0;
 cl_command_queue commandQueue = 0;
 cl_program program = 0;
 cl_device_id device = 0;
 cl_kernel kernel = 0;
 cl_mem memObjects[3] = { 0, 0, 0 };
 cl_int errNum;
 //创建OpenCL上下文
 context = CreateContext(&device);
 if (context == NULL)
 {
    printf("Failed to create OpenCL context." );
    return 1;
}
 //获得OpenCL设备,并创建命令队列
 commandQueue = CreateCommandQueue(context, &device);
 if (commandQueue == NULL)
 {
    Cleanup(context, commandQueue, program, kernel,
              memObjects);
    return 1;
 }
 //创建OPenCL程序
 program = CreateProgram(context, device, "vecAdd.cl");
 if (program == NULL)
 {
    Cleanup(context, commandQueue, program,
              kernel, memObjects);
    return 1;
 }
 //创建OpenCL内核
 kernel = clCreateKernel(program, "vector_add", NULL);
 if (kernel == NULL)
 {
    printf( "Failed to create kernel");
    Cleanup(context, commandQueue, program,
              kernel, memObjects);
    return 1;
 }
 //创建OpenCL内存对象
 float result[ARRAY_SIZE];
 float a[ARRAY_SIZE];
 float b[ARRAY_SIZE];
 for (int i = 0; i 〈 ARRAY_SIZE; i++)
 {
    a[i] = (float)i;
    b[i] = (float)(i * 2);
 }
 if (!CreateMemObjects(context, memObjects, a, b))
 {
    Cleanup(context, commandQueue, program,
              kernel, memObjects);
    return 1;
 }
 //设置内核参数
 errNum = clSetKernelArg(kernel, 0, sizeof(cl_mem), &memObjects[0]);
 errNum |= clSetKernelArg(kernel, 1, sizeof(cl_mem),
&memObjects[1]);
 errNum |= clSetKernelArg(kernel, 2, sizeof(cl_mem),
                            &memObjects[2]);
 if (errNum != CL_SUCCESS)
 {
    printf("Error setting kernel arguments.");
    Cleanup(context, commandQueue, program,
              kernel, memObjects);
    return 1;
 }
 size_t globalWorkSize[1] = { ARRAY_SIZE };
 size_t localWorkSize[1] = { 1 };
 //执行内核
 errNum = clEnqueueNDRangeKernel(commandQueue, kernel,
                                      1, NULL,
                                      globalWorkSize,
                                      localWorkSize, 0, NULL,
                                      NULL);
 if (errNum != CL_SUCCESS)
 {
    printf( "Error queuing kernel for execution." );
    Cleanup(context, commandQueue, program, kernel,
            memObjects);
    return 1;
 }
 //计算结果拷贝回主机
 errNum = clEnqueueReadBuffer(commandQueue, memObjects[2],
                                  CL_TRUE, 0,
                                  ARRAY_SIZE * sizeof(float),
                                  result, 0, NULL, NULL);
 if (errNum != CL_SUCCESS)
 {
    printf( "Error reading result buffer." );
    Cleanup(context, commandQueue, program, kernel,
              memObjects);
    return 1;
 }
 for (int i = 0; i 〈 ARRAY_SIZE; i++)
 {
    printf("i=%d:%f\n", i, result[i]);
 }
 printf("Executed program succesfully." );
 Cleanup(context, commandQueue, program, kernel, memObjects);
 return 0;
}



我们以向量加法作为OpenCL编程的第一个例子。从例子中可以看出OpenCL编程的标准开发流程。

支持OpenCL支持的平台、设备很多,为了兼顾不同设备,OpenCL程序的第一步就是确定OpenCL执行的平台,在确定平台之后再确定执行OpenCL计算的设备。确定设备后创建上下文,上下文中包含上一步查询的OpenCL计算的设备(允许包含多个计算设备对象),以及接下来创建的内核、程序对象和内存对象。

创建上下文以后,需要创建命令队列,但一个OpenCL设备可以对应多个队列对应一个命令队列一个OpenCL设备。例如,如果需要使用上下文中包含的两个计算设备时,为每个设备创建各自的命令队列。主机与OpenCL设备间数据传输、执行内核等交互操作都是通过入队到命令队列中,命令队列中的各个命令交给OpenCL驱动或相应的硬件单元去执行。

由于在运行时才知道OpenCL设备信息,所以在OpenCL主机端程序中读取内核源码并创建程序对象,根据OpenCL设备编译、构建程序对象,最后创建内核对象。通过这三步操作,把OpenCL设备上执行的内核代码编译完成。

对于在OpenCL设备上执行的内核函数需要输入参数,在主机端调用设置内核参数函数。除此之外,还需要设置在设备上用于执行的工作组和工作项的参数。

当上述操作都完成以后,我们就可以把内核函数入队到命令队列中,命令队列提交给相应的设备执行。

OpenCL设备执行完计算后,把数据拷贝回主机端,并销毁分配的资源。

上述整个流程,如下图所示。

OpenCL执行流程

通过向量加法的例子,相信读者对于OpenCL开发有了一定的了解。

赞(1)
未经允许不得转载:极客笔记 » OpenCL 第一个程序

评论 抢沙发

  • 昵称 (必填)
  • 邮箱 (必填)
  • 网址
OpenCL 基本概念
OpenCL 是什么OpenCL 平台模型OpenCL 执行模型OpenCL 上下文简介OpenCL 命令队列简介OpenCL 在设备上执行内核OpenCL 存储器区域OpenCL 存储器对象OpenCL 共享虚拟存储器OpenCL 与OpenGL
OpenCL 基础教程
OpenCL 在Windows上搭建开发环境OpenCL 在Linux上搭建开发环境OpenCL 在OS X上搭建开发环境OpenCL 第一个程序OpenCL 平台OpenCL 设备OpenCL 创建上下文OpenCL 创建命令队列OpenCL 创建Program对象OpenCL 编译Program对象OpenCL 查询和管理Program对象OpenCL 创建内核对象OpenCL 设置内核参数OpenCL 查询和管理内核对象OpenCL 执行内核OpenCL 编写内核代码OpenCL 错误处理
OpenCL C特性
OpenCL 地址空间修饰符OpenCL 函数修饰符OpenCL 对象访问修饰符OpenCL 标量数据类型OpenCL 为什么要有矢量数据类型OpenCL 矢量初始化OpenCL 读取和修改矢量分量OpenCL 运算符OpenCL 维度和工作项OpenCL 工作组OpenCL 矢量数据拷贝OpenCL 异步拷贝和预取OpenCL 数学函数OpenCL 公共函数OpenCL 几何函数OpenCL 整数函数OpenCL 关系函数OpenCL 杂项矢量函数OpenCL 同步函数OpenCL 原子函数OpenCL 内建图像读函数OpenCL 内建无采样器图像读函数OpenCL 内建图像写函数OpenCL 内建图像查询函数OpenCL 工作组函数OpenCL 内建管道读/写函数OpenCL 内建工作组管道读/写函数OpenCL 内建管道查询函数OpenCL 设备队列OpenCL Blocks语法OpenCL 设备队列相关函数OpenCL 子内核存储器可见性OpenCL 设备队列的使用示例
OpenCL 存储器对象
OpenCL 存储器对象OpenCL 分配缓冲区对象OpenCL 创建子缓冲区对象OpenCL 图像对象和采样器对象OpenCL 图像对象OpenCL 图像格式描述符OpenCL 图像描述符OpenCL 图像对象查询OpenCL 采样器对象OpenCL 主机端采样器对象OpenCL 设备端采样器对象OpenCL 图像旋转示例OpenCL 管道OpenCL 创建管道对象OpenCL 管道对象查询OpenCL 主机与设备间数据传输OpenCL 图像对象主机与设备间数据拷贝OpenCL 缓冲区对象数据填充OpenCL 图像对象数据填充OpenCL 缓冲区对象间数据传输OpenCL 图像对象和缓冲区对象间数据拷贝OpenCL 缓冲区对象映射OpenCL 图像对象映射OpenCL 解映射OpenCL 共享虚拟存储器OpenCL SVM缓冲创建与释放OpenCL SVM缓冲映射与解映射OpenCL SVM缓冲填充与拷贝OpenCL SVM类型OpenCL SVM特性OpenCL 共享虚拟存储器示例OpenCL 存储器一致性模型OpenCL 存储器次序规则OpenCL 原子操作的存储器次序规则OpenCL 栅栏操作的存储器次序规则OpenCL 工作组函数的存储器次序规则OpenCL 主机端与设备端命令的存储器次序规则OpenCL 关于存储器次序在实际OpenCL计算设备中的实现
OpenCL 同步及事件机制
OpenCL 同步及事件机制OpenCL 主机端的OpenCL同步OpenCL OpenCL事件机制OpenCL 对OpenCL事件的标记和栅栏OpenCL 内核程序中的同步OpenCL 工作组内同步OpenCL 原子操作OpenCL 1.2中的原子操作OpenCL 2.0中的原子操作OpenCL 局部存储器与全局存储器间的异步拷贝OpenCL 工作组间同步
OpenCL 与OpenGL互操作
OpenCL 与OpenGL互操作OpenCL 从一个OpenGL上下文来创建OpenCL上下文OpenCL 使用OpenGL共享的缓存对象OpenCL 使用OpenGL纹理数据OpenCL 共享OpenGL渲染缓存OpenCL 从一个OpenCL存储器对象查询OpenGL对象信息OpenCL 访问共享对象的OpenCL与OpenGL之间的同步OpenCL AMD Cayman架构GPUOpenCL AMD GCN架构的GPUOpenCL NVIDIA CUDA兼容的GPUOpenCL NVIDIA GPU架构的执行模型OpenCL NVIDIA GPU的全局存储器OpenCL NVIDIA GPU的局部存储器OpenCL ARM Mali GPU硬件架构OpenCL ARM Mali GPU存储器层次OpenCL ARM Mali GPU OpenCL映射