OpenCL 访问共享对象的OpenCL与OpenGL之间的同步

为了要确保共享对象的数据完整性,OpenCL扩展提供了CL与GL对象访问的同步机制。例如,当我们要从一个OpenGL纹理对象来创建一个OpenCL图像类型的存储器对象时,倘若OpenGL的纹理数据没有生成完,那么OpenCL那边所获得的图像数据将是不完整的。我们尽管可以通过在OpenGL上下文中调用glFinish这种同步API,但是效率非常低!因为此API与clFinish()语义差不多,要将当前OpenGL所有命令都执行完后才能让当前主机端的线程继续执行下去,否则当前线程将会被阻塞。这会导致很多不必要的计算资源的浪费。所以,OpenCL扩展API中提供了clEnqueueAcquireGLObjects这个函数,以获得相应的一个OpenCL事件对象。在OpenCL上下文中可以直接用此事件对象来做相应的OpenCL命令执行的同步就能更高效地利用计算资源了。下面介绍一下这个函数:

cl_int clEnqueueAcquireGLObjects(cl_command_queue command_queue,
                                        cl_uint num_objects.
                                        const cl_mem *mem_objects,
                                        cl_uint num_events_in_wait_list,
                                        const cl_event *event_wait_list,
                                        cl_event *event)
  • 参数num_objects指定了参数mem_objects所指向的一维数组中包含的所要获得的从GL对象所创建出的OpenCL存储器对象的个数。
  • 参数mem_objects就指向一个包含所需要获得的OpenCL存储器对象的数组首地址。
    后三个参数与其他clEnqueue系API的一样。

当这个函数调用成功后会返回CL_SUCCESS,如果是其他值说明调用失败。另外,我们往往需要最后一个参数event所生成的事件对象来追踪mem_objects中的所有OpenCL存储器对象的数据是否都在OpenGL上下文中生成完毕。所以,我们需要利用该事件对象对其他排入当前命令队列的命令进行同步。
在调用了clEnqueueAcquireGLObjects之后,必须在OpenGL上下文中再次使用OpenCL存储器对象所绑定的OpenGL对象之前调用clEnqueueReleaseGLObjects函数来释放所获得的OpenCL对象。该函数原型为:

cl_int clEnqueueReleaseGLObjects (cl_command_queue command_queue,
cl_uint num_objects,
                                          const cl_mem *mem_objects,
                                          cl_uint num_events_in_wait_list,
                                          const cl_event *event_wait_list,
                                          cl_event *event)

该函数参数与clEnqueueAcquireGLObjects的一模一样。我们在调用此函数时,所传递的num_objects和mem_objects参数也要同调用clEnqueueAcquireGLObjects时完全一致。
下面就用上面的核心代码稍作修改来给出使用clEnqueueAcquireGLObjects和clEnqueue-ReleaseGLObjects的例子。



- (void)generateTextureWithCLMemFromRBO
{
    /* 做OpenCL初始化 */
    cl_platform_id oclPlatform = NULL;
    cl_device_id oclDevice = NULL;
    //要被创建的OpenCL上下文对象
    cl_context context = NULL;
    cl_command_queue commandQueue = NULL;
    cl_program oclProgram = NULL;
cl_kernel kernel = NULL;
    //从OpenGL的RBO创建用于输出的OpenCL存储器对象
    cl_mem dstMem = NULL;
    //从OpenGL的RBO创建用于输入的OpenCL存储器对象
    cl_mem srcMem = NULL;
    //访问图像对象的采样器
    cl_sampler sampler = NULL;
#ifdef _APPLE_
    CGLContextObj cgl_context = CGLGetCurrentContext();
    CGLShareGroupObj sharegroup = CGLGetShareGroup(cgl_context);
    gcl_gl_set_sharegroup(sharegroup);
#endif
    do
    {
        //获得当前OpenCL平台
        cl_int status = clGetPlatformIDs(1, &oclPlatform, NULL);
        if(status != CL_SUCCESS)
        {
              NSLog(@"OpenCL platform get failed!");
              break;
        }
        //获得当前GPU设备。严格地来说,
        //此GPU设备也应该是OpenGL所使用的设备
        status = clGetDeviceIDs(oclPlatform, CL_DEVICE_TYPE_GPU, 1,
                                      &oclDevice, NULL);
        if(status != CL_SUCCESS)
        {
              NSLog(@"OpenCL GPU cannot be found!");
              break;
        }
        //设置用于创建OpenCL上下文的属性列表
        cl_context_properties properties[] =
        {
#ifdef WIN32
            CL_GL_CONTEXT_KHR ,
            (cl_context_properties)wglGetCurrentContext(),
            CL_WGL_HDC_KHR ,
            (cl_context_properties)wglGetCurrentDC(),
#endif
#ifdef _linux_
            CL_GL_CONTEXT_KHR ,
            (cl_context_properties)glXGetCurrentContext(),
            CL_GLX_DISPLAY_KHR ,
            (cl_context_properties)glXGetCurrentDisplay(),
#endif
#ifdef _APPLE_
            CL_CONTEXT_PROPERTY_USE_CGL_SHAREGROUP_APPLE,
            (cl_context_properties)sharegroup,
#endif
            0
        };
        //创建OpenCL上下文
        context = clCreateContext(properties, 1, &oclDevice,
                                          NULL, NULL, NULL);
        //创建命令队列
        commandQueue = clCreateCommandQueue(context, oclDevice,
                                                              0, NULL);
        //编译内核程序
        NSString *kernelPath = [[NSBundle mainBundle]
                                          pathForResource:
                            @"compute" ofType:@"ocl"];
        const char *aSource = [[NSString stringWithContentsOfFile:
                            kernelPath
            encoding:NSUTF8StringEncoding error:nil] UTF8String];
        size_t kernelLength = strlen(aSource);
        oclProgram = clCreateProgramWithSource(context, 1, &aSource,
                                                      &kernelLength, NULL);
        if(oclProgram == NULL)
        {
            NSLog(@"OpenCL program create failed!");
            break;
        }
        //构建程序
        status = clBuildProgram(oclProgram, 1, &oclDevice,
                                  NULL, NULL, NULL);
if(status != CL_SUCCESS)
        {
            NSLog(@"OpenCL kernel build failed!");
            break;
        }
        //创建与GL纹理对象共享的CL图像存储器对象,用于纹理输出
        dstMem = clCreateFromGLTexture(context, CL_MEM_WRITE_ONLY,
                                        GL_TEXTURE_2D, 0, mTexName,
                                        NULL);
        //从GL渲染缓存创建与CL共享的存储器对象,作为输入源
        srcMem = clCreateFromGLRenderbuffer(context,
                                              CL_MEM_READ_ONLY,
                                              mRBO, NULL);
        //创建采样器对象
        sampler = clCreateSampler(context, CL_FALSE,
                                  CL_ADDRESS_CLAMP_TO_EDGE,
                                  CL_FILTER_LINEAR, NULL);
        //创建内核对象
        kernel = clCreateKernel(oclProgram, "ImageProcessing",
                              NULL);
        //设置内核参数
        status = clSetKernelArg(kernel, 0, sizeof(dstMem), &dstMem);
        status |= clSetKernelArg(kernel, 1, sizeof(srcMem),
                                &srcMem);
        status |= clSetKernelArg(kernel, 2, sizeof(cl_sampler),
                                (void *)&sampler);
        if(status != CL_SUCCESS)
        {
            NSLog(@"Kernel parameters pass failed!");
            break;
        }
        //用于跟踪获得GL对象的事件
        cl_event acquireEvt = NULL;
        //获得从GL对象所创建的CL对象,等待GL绘制完成
        clEnqueueAcquireGLObjects(commandQueue, 2, (cl_mem[])
        {
            dstMem, srcMem
        }, 0, NULL, &acquireEvt);
        //这里我们总共使用mImageWidth * mImageHeight个工作项,
        //每个工作项来处理一个像素
        size_t global_work_size[] = { mRenderWidth, mRenderHeight };
        //然后设置一个工作组中的工作项个数。
        //要注意,x维度与y维度两个数相乘
        //不能大于工作组中最多可容纳的工作项的个数
        size_t local_work_size[] = { 16, 16 };
        //运行内核程序
        status |= clEnqueueNDRangeKernel(commandQueue, kernel, 2,
                                          NULL, global_work_size,
                      local_work_size, 1, &acquireEvt, NULL);
        //这里直接用clFinish进行同步,确保顶点坐标以及相应的颜色值全都设置好
        clFinish(commandQueue);
        //调用过clEnqueueAcquireGLObjects之后必须调用此函数,
        //并且在后续使用相关OpenGL对象之前完成调用
        clEnqueueReleaseGLObjects(commandQueue, 2, (cl_mem[]){
                                      dstMem, srcMem
                                      }, 0, NULL, NULL);
        clReleaseEvent(acquireEvt);
        if(status != CL_SUCCESS)
        {
            NSLog(@"OpenCL kernel run error!");
        }
    }
    while(NO);
    //释放OpenCL各种对象
    if(dstMem != NULL)
        clReleaseMemObject(dstMem);
    if(srcMem != NULL)
        clReleaseMemObject(srcMem);
    if(sampler != NULL)
        clReleaseSampler(sampler);
    if(kernel != NULL)
        clReleaseKernel(kernel);
    if(oclProgram != NULL)
        clReleaseProgram(oclProgram);
    if(commandQueue != NULL)
clReleaseCommandQueue(commandQueue);
    if(context != NULL)
        clReleaseContext(context);
}
- (void)drawRect:(NSRect)dirtyRect
{
    //绑定设置好的帧缓存对象与渲染缓存对象
    glBindFramebuffer(GL_FRAMEBUFFER, mFBO);
    glBindRenderbuffer(GL_RENDERBUFFER, mRBO);
    //准备在绑定的渲染缓存上绘制彩色的圆
    [self doOpenCLComputing];
    glClear(GL_COLOR_BUFFER_BIT);
    glDrawArrays(GL_TRIANGLE_FAN, 0, 362);
    //这里将使用acquire接口进行同步,而不直接使用glFinish()
    //glFinish();
    //准备将渲染缓存中的图像变为纹理,然后交给OpenCL做进一步的处理
    [self prepareForDraw];
    [self generateTextureWithCLMemFromRBO];
    //使用默认帧缓存与渲染缓存,准备绘制
    glBindFramebuffer(GL_FRAMEBUFFER, 0);
    glBindRenderbuffer(GL_RENDERBUFFER, 0);
    glClear(GL_COLOR_BUFFER_BIT);
    glDrawArrays(GL_TRIANGLE_STRIP, 0, 4);
    glFlush();
    [[self openGLContext] flushBuffer];
}

在drawRect方法中,我们看到之前的glFinish()的调用被注释掉了,取而代之的就是在generateTextureWithCLMemFromRBO方法中使用clEnqueueAcquireGLObjects进行同步。通过该函数所返回的acquireEvt事件对象作为clEnqueueNDRangeKernel的同步事件。在完成dstMem与srcMem对象的数据生成之前,OpenCL内核命令的执行会一直被阻塞。在内核命令执行完之后,我们即可调用clEnqueueReleaseGLObjects来释放与OpenGL对象相关联的OpenCL存储器对象。紧接着generateTextureWithCLMemFromRBO方法下面,我们就开始做OpenGL端的图形绘制。这时,OpenCL端的工作全部完成,且OpenCL上下文及各种存储器对象也可以全部销毁。

赞(2)
未经允许不得转载:极客笔记 » OpenCL 访问共享对象的OpenCL与OpenGL之间的同步

评论 抢沙发

  • 昵称 (必填)
  • 邮箱 (必填)
  • 网址
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映射