PowerVR Rogue图形处理器内核编写快速指南(一)

本文及下月发布的后续文章会介绍关于PowerVRRogue系列架构的OpenCL编程。

首先,我会向你展示OpenCL的编程基础(其使用基本程序),和Rogue图形处理器的语言执行说明。

接下来我会用图像过滤程序案例来说明Rogue构架的编程指南。

OpenCL概述

设想一个简单的C程序,使两个二维阵列的对应元素相乘:
voidmatrix_mul_cpu(const float *a, const float *b, float *c, int l, int h)
{
inti;
for (i=0; i for (j=0; j c[i][j] = a[i][j] * b[i][j];
}

该for循环由两部分组成:向量(包含l*h个元素的二维向量)及乘法运算。各个循环迭代相互独立,之间不存在数据相关。这意味着这些乘法运算可在GPU的无数线程中并行运算:

OpenCL程序由两部分构成:一个在主机CPU上运行,另一个在图形处理器上运行。要想在GPU上执行代码,首先要定义OpenCL内核,内核由含有额外关键字及数据类型的变体C语言写成。接着前面的案例,一个OpenCL内核实现简单乘法的方式如下:
_kernel void vector_mul(__global const float *a, __global const float *b, __global float *c)
{
size_ti = get_global_id(0);
size_t j = get_global_id(1);
c[i][j] = a[i][j] * b[i][j];
}

主机程序启动(或利用)这个内核,生成一个大小为512(32×16)并行工作项的虚拟网格,如下所示。网格是N维向量,其中N可取1,、2或3。单个工作项执行N维向量的一个运算点。执行相同核代码的工作项具有各自不同的ID,通过ID它们可以被访问和调用数据。此ID可以通过内置函数get_global_id获取。

OpenCL

在OpenCL中,多个工作项组成工作组。上图中,每个工作组由8×4共32个工作项组成。同一工作组的工作项之间通过本地内存同步和共享数据(后文有解释)。更多精彩内容,请点击这里

--电子创新网--
粤ICP备12070055号