Test case attached, and reproduced below. The last one or two arguments to foo are always zero for some reason, regardless of the number of arguments. The number of valid arguments appears to always be even. Removing the initialisation of the returned struct causes the bug to go away.
typedef struct {
int bar;
uint2 baz;
} Foo;
Foo foo_new(int offset, int offset2, int offset3, int offset4, int o5, int o6) {
Foo foo = {
.bar = 1,
.baz = {2, 3},
};
printf("Received arguments: %d, %d, %d, %d, %d, %d\n", offset, offset2, offset3, offset4, o5, o6);
return foo;
}
__kernel void testcase() {
for (int i = 0; i < 1000; i++) {
Foo foo = foo_new(i, i, i, i, i, i);
}
}
Thanks for the testcase. I will let you know, if i am able to reproduce it.
I am not able to see it at my end. Can you check with Catalyst 13.6 beta.
Also let us know the system details you are currently working : CPU, GPU, Catalyst, SDK, OS.
After some testing, it appears to only occur with -cl-opt-disable. Also, when run on the GPU instead of the CPU, it hard locks the system.
CPU: Intel(R) Core(TM) i5-3570 CPU @ 3.40GHz
GPU: AMD Radeon HD 7970
Catalyst: 13.6 beta
OpenCL version: OpenCL 1.2 AMD-APP (1214.3)
OS: Funtoo Linux 64bit
Reproduced for both CPU & GPU. Forwarding it to AMD Engg team.
We were unable to reproduce this issue with the latest driver. Can you verify and confirm?