linux | 苹果OpenCL(提高应用软件如游戏、娱乐以及科研和医疗软件的运行速度和响应)

点击上方"蓝字"关注我们

01、引言

>>>

OpenCL 1.0 于 2008 年 11 月发布。

OpenCL 是为个人电脑、服务器、移动设备以及嵌入式设备的多核系统提供并行编程开发的底层 API。OpenCL 的编程语言类似于 C 语言。其可以用于包含 CPU、GPU 以及来自主流制造商如 NXP®、NVIDIA®、Intel®、AMD、IBM 等的处理器的异构平台。OpenCL 旨在提高应用软件如游戏、娱乐以及科研和医疗软件的运行速度和响应。

我们使用 Apalis iMX6Q 系统模块测试 OpenCL，对比两个应用 - 一个运行在 GPU 上，另一个则在 CPU。最后我们将分享本次测试的结果。

02、Toradex Embedded Linux 镜像中添加 OpenCL

>>>

假设你已经具有能够编译 Apalis iMX6 镜像的 OpenEmbedded 编译环境。你可以参考我们 OpenEmbedded (core) 【http://developer.toradex.com/knowledge-base/board-support-package/openembedded-(core)】文章。

为编译支持 OpenCL 以及相关库文件的嵌入式 Linux 镜像，需要采取以下步骤：

首先，修改下面目录中的文件。

/meta-toradex/recipes-fsl/packagegroups/packagegroup-fsl-tools-gpu.bbappend

添加如下内容：

SOC_TOOLS_GPU_append_mx6 = " \libopencl-mx6 \libgles-mx6 \
"

并在 local.conf 文件中添加 imx-gpu-viv

IMAGE_INSTALL_append = "imx-gpu-viv"

现在就可以编译镜像：

bitbake angstrom-lxde-image

03、GPU 和 CPU 代码

>>>

所有的代码可以从 GitHub【https://github.com/giobauermeister/OpenCL-test-apps】上下载。

我们使用数列求和应用作为基本的演示例程。第一部分代码运行在 GPU 上，第二部分则在 CPU 上。应用执行完毕后打印其所消耗的时间。使用 OpenCL 所需的头文件是 cl.h，位于文件系统的 /usr/include/CL 目录。链接程序所需的库文件是 libGAL.so 和 libOpenCL.so，位于 /usr/lib 目录。

为了计算消耗的时间，我们创建带分析功能的队列，在结束的时候获取分析的结果。

下面是 OpenCL 代码：

//************************************************************
// Demo OpenCL application to compute a simple vector addition
// computation between 2 arrays on the GPU
// ************************************************************
#include 
#include 
#include 
#include <CL/cl.h>
//
// OpenCL source code
const char* OpenCLSource[] = {
"__kernel void VectorAdd(__global int* c, __global int* a,__global int* b)",
"{",
" // Index of the elements to add \n",
" unsigned int n = get_global_id(0);",
" // Sum the nth element of vectors a and b and store in c \n",
" c[n] = a[n] + b[n];",
"}"
};
// Some interesting data for the vectors
Int InitialData1[80] = {37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17};
int InitialData2[80] = {35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15};
// Number of elements in the vectors to be added
#define SIZE 600000
// Main function
// ************************************************************
int main(int argc, char **argv)
{ // Two integer source vectors in Host memoryint HostVector1[SIZE], HostVector2[SIZE];//Output Vectorint HostOutputVector[SIZE];// Initialize with some interesting repeating datafor(int c = 0; c < SIZE; c++){HostVector1[c] = InitialData1[c%20];HostVector2[c] = InitialData2[c%20];HostOutputVector[c] = 0;}//Get an OpenCL platformcl_platform_id cpPlatform;clGetPlatformIDs(1, &amp;cpPlatform, NULL);// Get a GPU devicecl_device_id cdDevice;clGetDeviceIDs(cpPlatform, CL_DEVICE_TYPE_GPU, 1, &amp;cdDevice, NULL);char cBuffer[1024];clGetDeviceInfo(cdDevice, CL_DEVICE_NAME, sizeof(cBuffer), &amp;cBuffer, NULL);printf("CL_DEVICE_NAME: %s\n", cBuffer);clGetDeviceInfo(cdDevice, CL_DRIVER_VERSION, sizeof(cBuffer), &amp;cBuffer, NULL);printf("CL_DRIVER_VERSION: %s\n\n", cBuffer);// Create a context to run OpenCL enabled GPUcl_context GPUContext = clCreateContextFromType(0, CL_DEVICE_TYPE_GPU, NULL, NULL, NULL);     // Create a command-queue on the GPU devicecl_command_queue cqCommandQueue = clCreateCommandQueue(GPUContext, cdDevice, CL_QUEUE_PROFILING_ENABLE, NULL);// Allocate GPU memory for source vectors AND initialize from CPU memorycl_mem GPUVector1 = clCreateBuffer(GPUContext, CL_MEM_READ_ONLY |CL_MEM_COPY_HOST_PTR, sizeof(int) * SIZE, HostVector1, NULL);cl_mem GPUVector2 = clCreateBuffer(GPUContext, CL_MEM_READ_ONLY |CL_MEM_COPY_HOST_PTR, sizeof(int) * SIZE, HostVector2, NULL);// Allocate output memory on GPUcl_mem GPUOutputVector = clCreateBuffer(GPUContext, CL_MEM_WRITE_ONLY,sizeof(int) * SIZE, NULL, NULL);// Create OpenCL program with source codecl_program OpenCLProgram = clCreateProgramWithSource(GPUContext, 7, OpenCLSource, NULL, NULL);// Build the program (OpenCL JIT compilation)clBuildProgram(OpenCLProgram, 0, NULL, NULL, NULL, NULL);// Create a handle to the compiled OpenCL function (Kernel)cl_kernel OpenCLVectorAdd = clCreateKernel(OpenCLProgram, "VectorAdd", NULL);// In the next step we associate the GPU memory with the Kernel argumentsclSetKernelArg(OpenCLVectorAdd, 0, sizeof(cl_mem), (void*)&amp;GPUOutputVector);clSetKernelArg(OpenCLVectorAdd, 1, sizeof(cl_mem), (void*)&amp;GPUVector1);clSetKernelArg(OpenCLVectorAdd, 2, sizeof(cl_mem), (void*)&amp;GPUVector2);//create eventcl_event event = clCreateUserEvent(GPUContext, NULL);// Launch the Kernel on the GPU// This kernel only uses global datasize_t WorkSize[1] = {SIZE}; // one dimensional RangeclEnqueueNDRangeKernel(cqCommandQueue, OpenCLVectorAdd, 1, NULL, WorkSize, NULL, 0, NULL, &amp;event);// Copy the output in GPU memory back to CPU memoryclEnqueueReadBuffer(cqCommandQueue, GPUOutputVector, CL_TRUE, 0,SIZE * sizeof(int), HostOutputVector, 0, NULL, NULL);// CleanupclReleaseKernel(OpenCLVectorAdd);clReleaseProgram(OpenCLProgram);clReleaseCommandQueue(cqCommandQueue);clReleaseContext(GPUContext);clReleaseMemObject(GPUVector1);clReleaseMemObject(GPUVector2);clReleaseMemObject(GPUOutputVector);    clWaitForEvents(1, &amp;event);cl_ulong start = 0, end = 0;double total_time;     clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_START, sizeof(cl_ulong), &amp;start, NULL);clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_END, sizeof(cl_ulong), &amp;end, NULL);total_time = end - start;     printf("\nExecution time in milliseconds = %0.3f ms", (total_time / 1000000.0) );printf("\nExecution time in seconds = %0.3f s\n\n", ((total_time / 1000000.0))/1000 );          return 0;
}

CPU 代码是简单的 C 程序，和上面一样计算同样的队列求和。为了计算消耗的时间，我们使用 time.h中的库。代码如下：

#include 
#include 
#include  int InitialData1[80] = {37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17,37,50,54,50,56,0,43,43,74,71,32,36,16,43,56,100,50,25,15,17};
int InitialData2[80] = {35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15,35,51,54,58,55,32,36,69,27,39,35,40,16,44,55,14,58,75,18,15};#define SIZE 600000int main(int argc, char **argv)
{
time_t start, stop;
clock_t ticks;time(&start);    
// Two integer source vectors in Host memory
int HostVector1[SIZE], HostVector2[SIZE];
//Output Vector
int HostOutputVector[SIZE];
// Initialize with some interesting repeating data
//int n;
for(int c = 0; c < SIZE; c++)
{
HostVector1[c] = InitialData1[c%20];
HostVector2[c] = InitialData2[c%20];
HostOutputVector[c] = 0;
}for(int i = 0; i < SIZE; i++)
{HostOutputVector[i] = HostVector1[i] + HostVector2[i];ticks = clock();
}     time(&amp;stop);printf("\nExecution time in miliseconds = %0.3f ms",((double)ticks/CLOCKS_PER_SEC)*1000);printf("\nExecution time in seconds = %0.3f s\n\n", (double)ticks/CLOCKS_PER_SEC);return 0;
}

04、交叉编译应用

>>>

同一个 Makefile 可以用于交叉编译 GPU 和 CPU 应用。你需要注意下面的三个变量。根据你的系统做相应的调整：

• ROOTFS_DIR -> Apalis iMX6 文件系统路径

• APPNAME -> 应用的名字

• TOOLCHAIN -> 交叉编译工具的路径

export ARCH=arm
export ROOTFS_DIR=/usr/local/toradex-linux-v2.5/oe-core/build/out-glibc/sysroots/apalis-imx6APPNAME = proc_sample
TOOLCHAIN = /home/prjs/toolchain/gcc-linaroCROSS_COMPILER = $(TOOLCHAIN)/bin/arm-linux-gnueabihf-
CC= $(CROSS_COMPILER)gcc
DEL_FILE = rm -rf
CP_FILE = cp -rf
TARGET_PATH_LIB = $(ROOTFS_DIR)/usr/lib
TARGET_PATH_INCLUDE = $(ROOTFS_DIR)/usr/include
CFLAGS = -DLINUX -DUSE_SOC_MX6 -Wall -std=c99 -O2 -fsigned-char -march=armv7-a -mfpu=neon -DEGL_API_FB -DGPU_TYPE_VIV -DGL_GLEXT_PROTOTYPES -DENABLE_GPU_RENDER_20 -I../include -I$(TARGET_PATH_INCLUDE)
LFLAGS = -Wl,--library-path=$(TARGET_PATH_LIB),-rpath-link=$(TARGET_PATH_LIB) -lm -lglib-2.0 -lOpenCL -lCLC -ldl -lpthread
OBJECTS = $(APPNAME).o
first: all
all: $(APPNAME)
$(APPNAME): $(OBJECTS)
$(CC) $(LFLAGS) -o $(APPNAME) $(OBJECTS)
$(APPNAME).o: $(APPNAME).c
$(CC) $(CFLAGS) -c -o $(APPNAME).o $(APPNAME).c
clean:
$(DEL_FILE) $(APPNAME)

在应用所在的目录中保持 Makefile 文件，然后运行 make。
将编译生成的文件复制到 Apalis iMX6 开发板上。

测试结果

在执行两个应用程序后，我们得到以下结果：

### Processor time
Execution time in miliseconds = 778.999 ms
Execution time in seconds = 0.779 s ### GPU time 
Execution time in milliseconds = 12.324 ms
Execution time in seconds = 0.012 s

根据以上结果，我们可以很清楚地看到在 Apalis iMX6Q GPU 上使用 OpenCL 能够加速队列求和运算。

总结

>>>

借助 OpenCL，可以在不同设备从图形显卡到超级计算机以及嵌入式设备，运行代码。用户还可以进一步结合，例如在 OpenCV 中使用 OpenCL 提高计算机视觉的性能。

对于绝大多数嵌入式应用，Linux 是正确的选择。Linux 编译系统，例如 Buildroot 和 OpenEmbedded，能够创建定制化的 BSP，裁剪到任意的大小，并且提供丰富的应用和 SDK，从 gstreamer、Python 到 node.js 等。基于 OpenEmbedded/Yocto 的 Linux 是 Toradex 支持的默认发行版本，开发社区还提供多种开发语言环境和框架。

现在的 GUI 可以使用 Qt、HTML5 来开发，以至于有点难于选择。

你也可以使用 NDK 基于 C/C++ 和 C# 开发你的应用，使用 Qt 作为显示框架或者利用 Cordova 或者 React Native 框架用 Javascript 开发移动应用

故我在

点击下方卡片 关注我

↓↓↓