全局设备变量 OpenCL

Posted

技术标签:

【中文标题】全局设备变量 OpenCL【英文标题】:Global device variable OpenCL 【发布时间】:2013-05-14 15:04:40 【问题描述】:

如何声明一个对 OpenCL 中的所有线程都是全局的设备变量? 我正在将一些代码从 CUDA 移植到 OpenCL。在我的 CUDA 实现中,我有类似的东西

...
...

__device__ int d_var;

...
...
void foo() 
   int h_var = 0
   cudaMemcpyToSymbol(d_var, h_var, sizeof(int));
   do
      //launch kernel, inside kernel d_var is modified
      cudaMemcpyFromSymbol(h_var, d_var, sizeof(int));
   while(h_var != 0);

我一直在阅读 OpenCL 示例代码,但不知道如何执行此操作。 任何建议都会很棒!

【问题讨论】:

【参考方案1】:

不幸的是,这并不是直接移植。虽然我不完全确定您是否至少可以在 OpenCL 程序中定义和使用(读/写)这样一个全局变量(但我不这么认为),但绝对没有办法访问(读/写)来自主机的这个变量。

你要做的就是把它作为一个额外的内核参数放入内核。如果它仅由主机写入并由内核读取,则普通的int 变量就足够了(因此每个内核都有一个副本)。但在您的情况下,因为它也是写入内核并从主机读取的,所以它必须是一个 __global 指针,并绑定了一个适当的缓冲区:

__kernel void func(..., __global int *d_var)  ... 

然后您可以使用通常的缓冲区功能在主机上读取和写入(当然,在内核中通过经典指针取消引用*d_var,但请记住,如果并发写入此变量会产生未指定的结果不是原子的)。

【讨论】:

抱歉,我的代码出错了。有一个主机 h_var 和一个设备 d_var,然后我从 device2host 来回复制。我想我将不得不按照您的建议创建另一个缓冲区。我尝试使用__kernel void foo(..., __constant int d_var),但出现错误,是因为 d_var 不是指针吗? @BRabbit27 “是因为 d_var 不是指针吗?” - 是的。我理解你的代码示例,答案仍然适用。您必须为您的 d_var 创建一个缓冲区(只是单个 int 大缓冲区,或较大缓冲区的子缓冲区)并从那里复制到 h_var,反之亦然,使用通常的缓冲区读/写操作。然后像往常一样将此缓冲区作为*d_var 参数的内核参数传递。顺便说一句,如果你想在内核中写信给d_var,就像你在问题中写的那样,你需要__global而不是__constant 我必须把它作为__global 放在内核中?如果我不输入预选赛怎么办?它转换为局部变量?两种方法都试过了,但都没有奏效。好吧,我的循环只运行一次,但我确信它必须运行更多次。 我又错了。我忘了取消引用内核内的指针。 *d_change = 1 我有它d_change = 1。工作完美:)【参考方案2】:

对于适合变量的小项目,将它们作为参数传递。

对于较大的常量项,使用 OpenCL“常量”地址空间。它针对广播使用进行了优化。

在您的 CUDA 示例中,您将数据读回。在 OpenCL 中,您需要为此使用常规缓冲区对象,因为不会将参数读回主机,并且常量内存是只读的。

【讨论】:

以上是关于全局设备变量 OpenCL的主要内容,如果未能解决你的问题,请参考以下文章

将全局内存用于(大)本地/私有温度。 OpenCL 中高效的数据结构

OpenCL 多 GPU 积分 - 将全局大小从 32 更改为 64 时的段错误

OpenCL 中的全局内存是不是连续

为啥 Cuda/OpenCL 的全局内存中没有银行冲突?

OpenCL - 全局内存读取性能优于本地

OpenCL 中的全局内存限制