浅谈编译(四)

gewenbin
gewenbin
gewenbin
188
文章
15
评论
2020年12月26日19:36:19 评论 595

OpenCL

1. OpenCL简介

OpenCL 是由 Khronos Group 针对异构计算设备(heterogeneous device)进行并行运算所设计的标准 API 以及程式语言。

OpenCL 程序分成两部分:一部分是在设备上执行的(例如 GPU),另一部分是在主机上(CPU)运行的。在设备上执行的程序就是实现“异构”和“并行计算”的部分。 为了能在设备上执行代码,程序员需要写一个特殊的程序(kernel 程序)。这个程序需要使用 OpenCL 语言编写。 OpenCL 语言采用了 C 语言的一部分加上一些约束、关键字和数据类型。在主机上运行的程序用 OpenCL 的 API 管理设备上运行的程序。主机程序的 API 用 C 语言编写,也有 C++、 Java、 Python 等高级语言接口。

设计OpenCL的目的主要有两个:

  • 并行计算
  • 同一份代码在不同的平台(异构)都能执行

OpenCL的根本目的就是针对并行计算提供一个统一的编程方式,从而方便对使用不同类型的并行计算硬件进行编程,这些硬件设备有多核CPU、GPU、FPGA、DSP、其他形式硬件等。如下所示:

浅谈编译(四)

2. Kernel程序

在OpenGL规范中,针对在GPU硬件上的编程提出了GLSL语言规范,同样的,在OpenCL规范中,针对在不同的并行计算设备上编程提出了“OpenCL C”语言规范。此语言设计参考了C语言,使用“OpenCL C”语言写的程序叫做kernel程序。通过一个简单的示例来直观的感受下如何在OpenCL程序中使用kernel程序。

void Init (void)
{
    char *kernelSource =
        "__kernel void test(__global int *pInOut)\n"
        "{\n"
        " int index = get_global_id(0);\n"
        " pInOut[index] += pInOut[index];\n"
        "}\n";
   
    cl_program program = NULL;
   
    // create program
    program = clCreateProgramWithSource(context, 1, (const char**)&kernelSource, NULL, &ret);

    // build program
    clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
}

假设以上的代码截取自demo.c源文件,对这个示例代码的说明如下:

  • kernelSource字符串表示的就是OpenCL kernel程序源码
  • 通过OpenCL接口clCreateProgramWithSource来加载kernel源码
  • kernel源码加载好之后需要进行编译,使用OpenCL接口clBuildProgram来实现这一目的

3. OpenCL驱动

OpenCL规范其实包含两个部分:

  • OpenCL API规范,使用OpenCL API编写的程序叫做OpenCL程序,比如上面的示例demo.c,是在CPU上运行的
  • OpenCL C规范,符合OpenCL C规范的程序叫做kernel程序,是在并行计算设备上运行的

OpenCL的驱动主要就有两大功能:

  • 实现OpenCL API
  • 实现OpenCL C编译器,用于将kernel程序编译成具体的并行计算设备硬件指令

如下图所示:

浅谈编译(四)

4. OpenCL编译器

这里主要讨论OpenCL 2.1版本之后的编译框架。OpenCL 2.1相比于之前的版本增加了两大功能:

  • 制定了OpenCL C++语言规范,之前的Kernel程序都是用OpenCL C语言编写的,现在可以使用OpenCL C++语言编写
  • OpenCL接口可以直接识别加载SPIR-V IR,然后编译成kernel程序在并行计算设备上执行

详细的框架如下所示:

浅谈编译(四)

从上图中可以看出以下几点信息:

  • 用OpenCL C++语言编写的kernel程序必须编译成SPIR-V格式的IR给OpenCL接口使用
  • 用OpenCL C语言编写的kernel程序可以直接传给OpenCL接口使用(主要是为了兼容考虑),也可以编译成SPIR-V格式的IR给OpenCL接口使用
  • OpenCL 2.1版本的驱动能处理OpenCL C和SPIR-V格式的kernel程序

从OpenCL 2.1版本的编译框架改动来看,Khronos组织希望使用SPIR-V这一IR来统一Vulkan和OpenCL的中间表示语言。Khronos提供了一个名为SPIR-V generator/Clang的基于LLVM的编译器,用于将用OpenCL C和OpenCL C++语言编写的kernel源码编译成SPIR-V格式的二进制文件。

gewenbin
  • 本文由 发表于 2020年12月26日19:36:19
  • 转载请务必保留本文链接:http://www.databusworld.cn/9825.html
匿名

发表评论

匿名网友 填写信息

:?: :razz: :sad: :evil: :!: :smile: :oops: :grin: :eek: :shock: :???: :cool: :lol: :mad: :twisted: :roll: :wink: :idea: :arrow: :neutral: :cry: :mrgreen: