Reputation: 415
I can't figure how to deal with arrays which size is smaller that the global size sets in clEnqueueNDRangeKernel
.
In my atomistic calculation the global size is set by the number of atom. This is the size of the position, velocities and forces. Then, I do a pseudo reduction and store the data in arrays of size corresponding to the number of blocs. I have also a small array to store the parameters of the pair potential.
I don't have any issue with the position, I declare them as __global const double4 *positions
. My problem is about the two other kind. From my understanding a thread is associated to each array element but that's not what I want for the small arrays. To me the best way to go is to pass the parameter array as __const
or __private
, but I'm not sure I can pass an array as __private
(like I do with scalar) and setting it as __const
produce the error:
:149:80: error: invalid address space for pointee of pointer argument to __kernel function
__global float2 *atom_type,__global const double *atmmass, __const double8 *two_body_type,
Regarding my reduce array, right now I'm passing it as __global
the output is just garbage. When I declare it as __const
I get the same error as above and, anyway I guess that would be an issue the get back the data afterward.
How do you guys would do?
Upvotes: 0
Views: 107
Reputation: 6333
The relationship between input or output arrays and the global size that your kernel executes over is not fixed or defined by OpenCL, it is completely up your kernel.
If you want all work items to work with larger array and then only have some work with smaller array, you can code that. Just check if the global_id is in the range of the smaller array.
If there is some fixed relationship between the array sizes, like 3 to 1, you might instead process there of the large array elements and just one of the small array elements per work item and use one third the global size.
Upvotes: 1