为什么OpenCL的Intel Kernel Builder告诉我我的内核没有矢量化？

2019-08-30 23:50:41 阅读：229 来源： 互联网

我是在一个有限的区域内编写一个内核来添加两个三维矩阵.
我有我的代码

#define PREC float

typedef struct _clParameter clParameter;
struct _clParameter {
    size_t width;
    size_t minWidth;
    size_t maxWidth;
    size_t height;
    size_t minHeight;
    size_t maxHeight;
    size_t depth;
    size_t minDepth;
    size_t maxDepth;
};

__kernel void clMatrixBasicOperate1Add(
    __global const PREC * restrict in1,
    __global const PREC * restrict in2,
    __global PREC * restrict out,
    __private const clParameter par) {

    size_t sizeOfXY = par.width * par.height;

    // 3-Dimension matrix

    size_t X = get_global_size(0);
    size_t x = get_global_id(0);

    size_t Y = get_global_size(1);
    size_t y = get_global_id(1);

    size_t Z = get_global_size(2);
    size_t z = get_global_id(2);

    size_t endX = (par.maxWidth - par.minWidth + 1)     / X;
    size_t endY = (par.maxHeight - par.minHeight + 1)   / Y;
    size_t endZ = (par.maxDepth - par.minDepth + 1)     / Z;

    if(x<( (par.maxWidth    - par.minWidth  + 1) % X) )     endX += 1;
    if(y<( (par.maxHeight   - par.minHeight + 1) % Y) )     endY += 1;
    if(z<( (par.maxDepth    - par.minDepth  + 1) % Z) )     endZ += 1;

    for(size_t k=0;k<endZ;k++)
    for(size_t j=0;j<endY;j++)
    for(size_t i=0;i<endX;i++) {
        size_t index = (par.minDepth + k*Z+z) * sizeOfXY + (par.minHeight + j*Y+y) * par.width + (par.  minWidth + i*X +x);
        out[index] = in1[index] + in2[index];
    }

    // return
}

当我使用Intel Kernel Builder For OpenCL API构建它时,它告诉我

Setting target instruction set architecture to: Default (Advanced Vector Extension (AVX))
OpenCL Intel CPU device was found!
Device name: Intel(R) Core(TM) i7-2630QM CPU @ 2.00GHz
Device version: OpenCL 1.2 (Build 83073)
Device vendor: Intel(R) Corporation
Device profile: FULL_PROFILE
Compilation started
Compilation done
Linking started
Linking done
Device build started
Device build done
Kernel <clMatrixBasicOperate1Add> was not vectorized
Done.
Build succeeded!

我想知道为什么clMatrixBasicOperate1Add没有vectorzied.

解决方法:

由于for循环中的终止条件,部分内核无法进行矢量化.这些条件都依赖于从内核输入计算的变量.因此,在内核编译时,英特尔OpenCL C编译器不知道这些循环将执行多少次迭代,因此根本无法优化它们.如果将内循环从for(size_t i = 0; i< endX; i)更改为for(size_t i = 0; i< 4; i),则内核得到矢量化.当然,这种改变并没有你想要的,但至少你的内核得到了矢量化:). 我认为您想要尝试的策略是沿着线程网格的X维度进行矢量化.这意味着您将沿X启动1/2线程数,但是使用vload2和vstore2函数来读取和写入全局内存.您也可以使用4,8或16个元素向量,在这种情况下,您将分别沿X维度启动当前线程数的1 / 4,1 / 8或1/16. 由于您使用的是第二代Core i7和浮点数据,因此您可能希望使用float8,vload8和vstore8,因为您的CPU支持同时运行8个浮点值的AVX指令.请注意,这不是性能可移植的,例如一些GPU可以很好地工作到float2,但使用float4 / 8/16时性能会下降.使用AMD CPU运行时的较旧CPU无法访问AVX指令,只能访问使用4元素浮点向量的SSE.因此,您应该使用像“-D vectype = float4”这样的字符串通过clBuildProgram选项中传递的宏来使矢量大小成为可调参数.

标签：c-3,linux,opencl,intel
来源： https://codeday.me/bug/20190830/1771655.html

本站声明： 1. iCode9 技术分享网（下文简称本站）提供的所有内容，仅供技术学习、探讨和分享；
2. 关于本站的所有留言、评论、转载及引用，纯属内容发起人的个人观点，与本站观点和立场无关；
3. 关于本站的所有言论和文字，纯属内容发起人的个人观点，与本站观点和立场无关；
4. 本站文章均是网友提供，不完全保证技术分享内容的完整性、准确性、时效性、风险性和版权归属；如您发现该文章侵犯了您的权益，可联系我们第一时间进行删除；
5. 本站为非盈利性的个人网站，所有内容不会用来进行牟利，也不会利用任何形式的广告来间接获益，纯粹是为了广大技术爱好者提供技术内容和技术思想的分享性交流网站。

ICode9

为什么OpenCL的Intel Kernel Builder告诉我我的内核没有矢量化？