全局设备变量 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 中高效的数据结构