OpenCL error: illegal implicit conversion between two pointers with different address spaces

Question

我有一个 float 函数 atomic_add:

inline void AtomicAdd(volatile __global float *source, const float operand)
{
    union{unsigned int intVal; float floatVal;} newVal;
    union{unsigned int intVal; float floatVal;} oldVal;

    do {
        oldVal.floatVal = *source;
        newVal.floatVal = oldVal.floatVal + operand;;
    }
    while (atomic_cmpxchg((volatile __global unsigned int *)source, oldVal.intVal, newVal.intVal) != oldVal.intVal);
}

这里是主内核，调用了上面的函数：

__kernel void main_kernel(__global float* image)
{
   AtomicAdd(&image[0], 1.0f); 
}

以上功能运行正确。现在我想引入一个局部变量来强制在一个工作组中一劳永逸地完成AtomicAdd：

__kernel void main_kernel(__global float* image)
{
   __local tmpSum[1];
   if(get_local_id(0)==0) { 
      tmpSum[0] = 0.0f; }

   barrier(CLK_LOCAL_MEM_FENCE);
   AtomicAdd(&tmpSum[0], 1.0f);
   barrier(CLK_LOCAL_MEM_FENCE);

   AtomicAdd(&image[0], tmpSum[0]); 
}

日志给出了以下错误：

Build log: 
:422:15: error: illegal implicit conversion between two pointers with         different address spaces
                            AtomicAdd(&tmpSum[0], 1.0f);
                                      ^~~~~~~~~~
:422:15: warning: passing 'float __attribute__((address_space(3))) *' to parameter of type 'float volatile __attribute__((address_space(1))) *' discards qualifiers
                            AtomicAdd(&tmpSum[0], 1.0f);
                                      ^~~~~~~~~~
:17:49: note: passing argument to parameter 'source' here
    inline void AtomicAdd(volatile __global float *source, const float operand)

如您所见，全局指针和局部指针之间的不同地址空间似乎有问题。有人知道这里发生了什么吗？

Answer 1

无法将 __local 指针传递给采用 __global 指针的函数，反之亦然。 AtomicAdd 函数的副本需要使用 __local 指针才能成功构建内核。

Answer 2

我把我的解决方案放在这里。正如 doqtor 所指出的，__local 与 _global 的地址 space 不同。所以最简单的方法是重载，不幸的是 opencl 不支持。我必须定义一个新的内联函数并解决问题。

OpenCL error: illegal implicit conversion between two pointers with different address spaces

OpenCL error: illegal implicit conversion between two pointers with different address spaces

atomic

local-variables

opencl