CODEWITHSUNDEEP
cudaopencl
BRabbit27
BRabbit27

Reputation: 6623

Global device variable OpenCL

How can I declare a device variable that is global to all the threads in OpenCL? I'm porting some code from CUDA to OpenCL. In my CUDA implementation I have something like

...
...

__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);
}

I've been reading through OpenCL example codes but cannot figure out how to do this. Any advise would be great !

Upvotes: 1

Views: 778

Answers (2)

Dithermaster
Dithermaster

Reputation: 6333

For small items that fit in a variable, pass them as arguments.

For larger constant items, use OpenCL "constant" address space. It is optimized for broadcast usage.

In your CUDA example, you read the data back out. In OpenCL you'd need to use a regular buffer object for this, since parameters are not read out back to the host, and constant memory is read-only.

Upvotes: 0

Christian Rau
Christian Rau

Reputation: 45948

Unfortunately this is not that straight-forward to port. Whereas I'm not completely sure if you can at least define and use (read/write) such a global variable in an OpenCL program (but I don't think so), there is definitely no way to access (read/write) this variable from the host.

What you will have to do is put it into the kernel as an additional kernel argument. If it is only writen by the host and read by the kernel an ordinary int variable would suffice (thus a copy for each kernel). But in your case as it is also written in the kernel and read from the host, it has to be a __global pointer, to which an appropriate buffer is bound:

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

Which you can then read from and write to on the host using the usual buffer functionality (and in the kernel through classic pointer dereference *d_var, of course, but keep in mind that concurrent writes to this variable have unspecified results if not atomic).

Upvotes: 1

Related Questions