OpenCL与GPU计算入门 译文

OpenCL与GPU计算入门

作者:	Erik Smistad
译者:	gashero
日期:	2015-05-26
标题原文:	Getting started with OpenCL and GPU Computing
地址:	http://www.thebigblob.com/getting-started-with-opencl-and-gpu-computing/

目录

• 1   安装和设置OpenCL
• 1.1   在Ubuntu Linux上安装AMD显卡上的OpenCL
• 1.2   在Ubuntu Linux上安装nVidia显卡上的OpenCL
• 2   第一个OpenCL程序，向量加法
• 2.1   内核
• 2.2   主机程序
• 3   编译OpenCL程序
• 4   学习更多

OpenCL(Open Computing Language，开放计算语言)是一个在并行多计算平台上写程序的框架，由多个制造商提供多种计算设备，如AMD、Intel、ATI、nVidia等。这个框架定义了一种语言来编写内核(kernel)。这些内核是一些运行在不同计算设备上的函数。本文解释如何开始OpenCL，以及两个列表并行求和的例子。

1   安装和设置OpenCL

最初需要下载最新的显卡驱动，这因为如果OpenCL在你的显卡驱动不支持OpenCL时无法工作。

要安装OpenCL，你需要下载一个OpenCL实现。主要的显卡制造商nVidia和AMD/ATI都有提供基于GPU的OpenCL实现。这些实现都提供了SDK和有用的工具，如性能优化器。下一步是下载和安装GPU SDK。注意不是所有的显卡都支持，支持的显卡参见制造商网站。

对AMD/ATI，下载 http://developer.amd.com/sdks/AMDAPPSDK/Pages/default.aspx 。

对nVidia，下载 http://developer.nvidia.com/object/cuda_download.html 。

安装步骤依赖于SDK和OS。遵循如下步骤。推荐使用Ubuntu Linux和AMD 7970显卡。

1.1   在Ubuntu Linux上安装AMD显卡上的OpenCL

要安装最新的AMD驱动到Ubuntu-12.04，需要安装和激活"ATI/AMD proprietary FGLRX显卡驱动"。

再这一步完成后，下载和解压 http://developer.amd.com/sdks/AMDAPPSDK/Pages/default.aspx 。

AMD APP SDK 2.8包含了安装器，运行:

sudo sh Install-AMD-APP.sh

下一步，安装OpenCL头文件:

sudo apt-get install opencl-headers

这就行了，注意AMD APP SDK，及其例子在 /opt/AMDAPP 。

1.2   在Ubuntu Linux上安装nVidia显卡上的OpenCL

下载CUDA SDK http://developer.nvidia.com/object/cuda_download.html 。打开终端运行安装文件:

sudo sh cudatoolkit_3.1_linux_64_ubuntu9.10.run

下载开发者驱动，安装前关闭X，运行文件然后开启X。其中关闭X用:

sudo /etc/init.d/gdm stop

然后打开终端通过Ctrl+Alt+F5，登录并导航到下载驱动的目录:

sudo sh devdriver_3.1_linux_64_256.40.run

之后重启X:

startx

在开始编译OpenCL应用前，需要添加库的路径:

export LD_LIBRARY_PATH=/usr/local/cuda/lib64

2   第一个OpenCL程序，向量加法

展示OpenCL的简单例子。假设有2个列表的数字，A和B，相同长度。向量加法就是将A和B对应的每个像素相加，并将结果放入新的列表C。

通过循环方式的如下:

for (int i=0; i<LIST_SIZE; i++) {

    C[i]=A[i]+B[i];

}

这个算法简单，但是是线性时间复杂度的，O(n)是列表大小。因为每次迭代都是独立的，因此操作是数据并行的，意味着每次迭代可以同时计算。所以我们有n个核心的处理器用以在常量时间内完成计算O(1)。

要编写OpenCL内核来并行计算。内核是一个函数，运行在计算设备上。

2.1   内核

内核以OpenCL语言编写，是C的子集，以及包含了很多数学和向量函数。内核做向量加法如下:

__kernel void vector_add(__global const int *A, __global const int *B, __global int *C) {

    //获得当前要处理元素的索引

    int i=get_global_id(0);

    //做计算

    C[i]=A[i]+B[i];

}

2.2   主机程序

主机程序控制了内核的执行。主机程序以C语言编写，但可以绑定到其他语言如C++和Python。OpenCL API定义于 cl.h (Apple为 opencl.h )。如下代码为主机程序执行了如上的内核。不会讲解细节，具体参见介绍性的 "The OpenCL Programming Book" 。主机程序的主要步骤：

1. 获取平台和设备的信息
2. 选择设备用以执行
3. 创建OpenCL上下文
4. 创建命令队列
5. 创建内存缓冲对象
6. 传输数据到设备的内存缓冲
7. 创建程序对象
8. 载入内核源码，并编译，或载入预编译的二进制OpenCL程序
9. 创建内核对象
10. 设置内核参数
11. 执行内核
12. 读取内存对象，此时从计算设备读取C列表

#include <stdio.h>

#include <stdlib.h>

 

#ifdef __APPLE__

#include <OpenCL/opencl.h>

#else

#include <CL/cl.h>

#endif

 

#define MAX_SOURCE_SIZE (0x100000)

 

int main(void) {

    //创建2个输入向量

    int i;

    cosnt int LIST_SIZE=1024;

    int *A=(int*)malloc(sizeof(int)*LIST_SIZE);

    int *B=(int*)malloc(sizeof(int)*LIST_SIZE);

    for (i=0; i<LIST_SIZE; i++) {

        A[i]=i;

        B[i]=LIST_SIZE-i;

    }

 

    //载入内核源码到source_str

    FILE *fp;

    char *source_str;

    size_t source_size;

 

    fp=fopen("vector_add_kernel.cl","r");

    if (!fp) {

        fprintf(stderr, "Failed to load kernel\n");

        exit(1);

    }

    source_str=(char*)malloc(MAX_SOURCE_SIZE);

    source_size=fread(source_str,1,MAX_SOURCE_SIZE,fp);

    fclose(fp);

 

    //获得平台和设备信息

    cl_platform_id platform_id=NULL;

    cl_device_id device_id=NULL;

    cl_uint ret_num_devices;

    cl_uint ret_num_platforms;

    cl_int ret=clGetPlatformIDs(1, &platform_id, &ret_num_platform);

    ret=clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_DEFAULT, 1, &deivce_id, &ret_num_devices);

 

    //创建OpenCL上下文

    cl_context context=clCreateContext(NULL,1,&device_id,NULL,NULL,&ret);

    //创建命令队列

    cl_command_queue command_queue=clCreateCommandQueue(context,device_id,0,&ret);

    //创建内存缓冲对象，在设备上为每个向量

    cl_mem a_mem_obj=clCreateBuffer(context,CL_MEM_READ_ONLY,

            LIST_SIZE*sizeof(int),NULL,&ret);

    cl_mem b_mem_obj=clCreateBuffer(context,CL_MEM_READ_ONLY,

            LIST_SIZE*sizeof(int),NULL,&ret);

    cl_mem c_mem_obj=clCreateBuffer(context,CL_MEM_WRITE_ONLY,

            LIST_SIZE*sizeof(int),NULL,&ret);

 

    //拷贝数据A和B到对应的内存缓冲

    ret=clEnqueueWriteBuffer(command_queue,a_mem_obj,CL_TRUE,0,

            LIST_SIZE*sizeof(int),A,0,NULL,NULL);

    ret=clEnqueueWriteBuffer(command_queue,a_mem_obj,CL_TRUE,0,

            LIST_SIZE*sizeof(int),B,0,NULL,NULL);

 

    //创建程序

    cl_program program=clCreateProgramWithSource(context,1,

            (const char**)&source_str,(const size_t*)&source_size, &ret);

    //构建程序

    ret=clBuildProgram(program,1,&device_id,NULL,NULL,NULL);

    //创建OpenCL内核

    cl_kernel kernel=clCreateKernel(program,"vector_add",&ret);

 

    //设置内核参数

    ret=clSetKernelArg(kernel,0,sizeof(cl_meme),(void*)&a_mem_obj);

    ret=clSetKernelArg(kernel,1,sizeof(cl_meme),(void*)&b_mem_obj);

    ret=clSetKernelArg(kernel,2,sizeof(cl_meme),(void*)&c_mem_obj);

 

    //执行内核

    size_t global_item_size=LIST_SIZE;  //处理整个列表

    size_t local_item_size=64;          //分割为64个组

    ret=clEnqueueNDRangeKernel(command_queue,kernel,1,NULL,

            &global_item_size,&local_item_size,0,NULL,NULL);

 

    //读取内存缓冲C到本地变量C

    int *C=(int*)malloc(sizeof(int)*LIST_SIZE);

    ret=clEnqueueReadBuffer(command_queue,c_mem_obj,CL_TRUE,0,

        LIST_SIZE*sizeof(int),C,0,NULL,NULL);

 

    //显示结果

    for (i=0; i<LIST_SIZE; i++) {

        printf("%d + %d = %d\n", A[i], B[i], C[i]);

    }

 

    //清理资源

    ret=clFlush(command_queue);

    ret=clFinish(command_queue);

    ret=clReleaseKernel(kernel);

    ret=clReleaseProgram(program);

    ret=clReleaseMemObj(a_mem_obj);

    ret=clReleaseMemObj(b_mem_obj);

    ret=clReleaseMemObj(c_mem_obj);

    ret=clReleaseCommandQueue(command_queue);

    ret=clReleaseContext(context);

    free(A);

    free(B);

    free(C);

    return 0;

}

想要OpenCL运行在GPU上，你需要修改常量 CL_DEVICE_TYPE_DEFAULT 到 CL_DEVICE_TYPE_GPU 。要运行在CPU上修改为 CL_DEVICE_TYPE_CPU 。所以修改很简单。

3   编译OpenCL程序

如果OpenCL头文件和库放在了正确的位置，如下命令编译向量加法程序:

gcc main.o -o vectorAddition -l OpenCL

4   学习更多

要学习更多关于OpenCL，建议从 "The OpenCL programming book" 开始。如下是其他的OpenCL信息：

1. nVidia's page on OpenCL： http://www.nvidia.com/object/cuda_opencl_new.html
2. AMD/ATI's page on OpenCL： http://developer.amd.com/zones/OpenCLZone/Pages/default.aspx
3. Official website of OpenCL： http://www.khronos.org/opencl/