在 OpenCL 内核中动态创建本地数组

2024-02-29

我有一个 OpenCL 内核,需要将一个数组作为多个数组进行处理,其中每个子数组总和都保存在本地缓存数组中。

例如,想象一下捕鸟数组:

[[1, 2, 3, 4], [10, 30, 1, 23]]
  • 每个工作组都有一个数组(在示例中我们有 2 个工作组);
  • 每个工作项处理两个数组索引(例如将值索引乘以 local_id),其中工作项结果保存在工作组共享数组中。

    __kernel void test(__global int **values, __global int *result, const int array_size){
        __local int cache[array_size];
    
        // initialise
        if (get_local_id(0) == 0){
            for (int i = 0; i < array_size; i++)
                cache[i] = 0;
        }
    
        barrier (CLK_LOCAL_MEM_FENCE);
    
        if(get_global_id(0) < 4){
            for (int i = 0; i<2; i++)
                cache[get_local_id(0)] += values[get_group_id(0)][i] * 
                                                             get_local_id(0);
        }
    
        barrier (CLK_LOCAL_MEM_FENCE);
    
        if(get_local_id(0) == 0){
            for (int i = 0; i<array_size; i++)
                result[get_group_id(0)] += cache[i];
        }
    }
    

问题是我无法使用内核参数定义缓存数组大小,但我需要这样做才能拥有动态内核。

我怎样才能动态创建它?就像c中的malloc函数一样...

或者唯一可用的解决方案是将临时数组发送到我的内核函数?


这可以通过添加来实现__local数组作为内核参数:

__kernel void test(__global int **values, __global int *result, 
    const int array_size, __local int * cache)

并提供所需的内核参数大小:

clSetKernelArg(kernel, 3, array_size*sizeof(int), NULL);

本地内存将在内核调用时分配。请注意,可能需要进行额外的检查以确保所需的本地内存大小不超过设备限制。

本文内容由网友自发贡献,版权归原作者所有,本站不承担相应法律责任。如您发现有涉嫌抄袭侵权的内容,请联系:hwhale#tublm.com(使用前将#替换为@)

在 OpenCL 内核中动态创建本地数组 的相关文章

随机推荐