Structure copy error Problem copy __global structure

Accelerated Computing CUDA CUDA Programming and Performance
1

I hope this is not a problem already reported. But there seems to be some PTX code generation problem with copying structures. The clBuildProgram returns errors with code that copies a structure on global address space to another.

The error looks like this:

ptxas ptx input, line 56; fatal : Parsing error near ‘.’: syntax error
: Retrieving binary for ‘anonymous_jit_identity’, for gpu=‘sm_13’, usage mode=’

The following is an example:

typedef struct
{
int x;
int y;
int z;
int w;
} SomeStructure;

__kernel void test_calc_kernel(__global void *heap)
{
__global SomeStructure *s1 = (__global SomeStructure *) heap;
__global SomeStructure *s2 = s1 + 1;
*s1 = *s2;
}

However, if I copy the structure to a __private structure as an intermediate step, it doesn’t complain. For example,

__kernel void test_calc_kernel(__global void *heap)
{
SomeStructure c;
__global SomeStructure *s1 = (__global SomeStructure *) heap;
__global SomeStructure *s2 = s1 + 1;
*s1 = c = *s2;
}

As far as I know, structure copy should be legal in OpenCL. Well, the workaround depends on that in any case case. But I may be wrong.

I am using CUDA 2.3 running on GTX 295 if that makes any difference. Thanks.