OpenCL 通道动态索引

Question

我想为矩阵乘法实现一个脉动结构。我的目标是为每个处理元素使用一个内核，因此我将从主机部分多次执行相同的内核。

为了在内核之间进行通信，我想使用通道或管道。问题是“频道扩展不支持对频道 ID 数组的动态索引”。内核的数量将取决于矩阵的大小，因此我需要一些方法来自动将通道连接到相应的内核。

总结一下，我正在寻找一种方法来创建这个功能：

channel float c0[32];

__kernel void producer (__global float * data_in){
    for(int i=0; i<32; i++){
        write_channel_altera(c0[i],data_in[i]);
    }
}

__kernel void consumer (__global float * ret_buf){
    for(int i=0; i<32; i++){
        ret_buf[i]=read_channel_altera(c0[i]);
    }
}

提前致谢！

Answer 1

OpenCL 通道（英特尔 FPGA 扩展）不支持“真正的”动态 索引，但在大多数情况下，您可以通过以下方式解决此限制 使用switch 或#pragma unroll 方法：

switch 方法在Intel FPGA SDK for OpenCL Programming Guide 中描述：

channel int ch[WORKGROUP_SIZE];

__kernel void consumer() {
    int gid = get_global_id(0);
    int value;

    switch(gid)
    {
        case 0: value = read_channel_intel(ch[0]); break;
        case 1: value = read_channel_intel(ch[1]); break;
        case 2: value = read_channel_intel(ch[2]); break;
        case 3: value = read_channel_intel(ch[3]); break;
        //statements

        case WORKGROUP_SIZE-1:read_channel_intel(ch[WORKGROUP_SIZE-1]); break;
    }
}

如果你有一个频道循环，你也可以使用#pragma unroll：

__kernel void consumer() {
    int values[WORKGROUP_SIZE]
    #pragma unroll
    for (int i = 0; i < WORKGROUP_SIZE; ++i) {
        values[i] = read_channel_intel(ch[i]);
    }
}