Hello everybody,
I’ve been facing a strange problem with constant memory. Look at the header of my kernel:
__kernel void Lvo
(
__global int* Rt,
__constant int* St,
__constant int* Me,
__constant int* Ie,
__global int* sl,
__global char* sqs,
__constant int* Q,
__constant int* amount,
__global int* A,
__global int* B,
__global int* C,
__global int* D,
__global int* E,
__global int* F,
__global int* J,
__global int* try_me,
__global int* s,
__global int* ss,
__global int* i,
__global int* k
)
{
...
}
This works just fine, but when I try to use the Rt argument in constant memory (by putting __constant instead of __global) I sometimes get a segmentation fault, sometimes a malloc corruption and sometimes a double-linked list error.
My Rt argument has a 2KB size and is the biggest constant structure, so the size limit isn’t a problem. Besides, it is used 9 times inside the kernel (always being read).
Can anybody help me or at least suggest something? I am tired of trying on my own.
Thanks
Samuel
I’m not sure what we can say other than “it looks like a bug in the compiler; please contact your vendor”.
The issue could be related to the number of __constant arguments rather than their size. Each literal value in the kernel source counts as one additional __constant argument and the limit is 8 per kernel IIRC.
Unfortunately this problem happens even if I put Rt as the only constant argument. You may be right, it looks like a bug in the compiler, but I’ll keep trying to find a way to solve this. Thanks.
Any other suggestions are welcome.
I still haven’t found the solution for my issue and searching on OpenCL reference guide I’ve found the following:
"A pointer to address space A can only be assigned to a pointer to the same address space A. Casting a pointer to address space A to a pointer to address space B is illegal. "
Does it mean that a variable allocated on private address space must not receive a value from a pointer to the constant address space?
Thank you
Samuel
Does it mean that a variable allocated on private address space must not receive a value from a pointer to the constant address space?
No, that’s not a problem. I’ll explain it with the following example:
__kernel void foo(__global int *a, __local int *b, __constant int *c)
{
// Illegal
a = b;
// Illegal
a = c;
// Illegal
b = c;
// Legal
a[0] = b[0];
// Legal
a[0] = c[0];
// Legal
b[0] = c[0];
}