I've found a critical bug/hang. I simply write to an int2 R/W buffer the values of a const struct and I add "one" to the buffer.
I've packed the source code and executables for you at
http://www.megaupload.com/?d=XT22I50T
It happens randomly so, please, run the executable several times until you get it ( usually 5-10 times from the MS-DOS console ). Usually the x86 version hangs faster than the x64.
I use Win7 x64, i7 920, Radeon 5750, Cat 10.12+SDK 2.3 normally.
Runs ok with Intel / NVIDIA OpenCL and Cat 10.9, so the problem should be a routine in the new Cat.10.10+ GPU code ( because the CPU one apparently runs ok ). Catalyst 10.10, 10.11 and 10.12 makes the hang to appear using GPU devices.
I'll put here the kernel too:
I've noticed in the SKA too a strange "LSHR" command. strange because I see several LSHL which I think are correct ( because the block/work sizes are pow2 ). That LSHR is probably computing bad the buffer offset and causing something to mess badly...
thx.
typedef struct tParams { int4 c1; int c21; int c22; int c23; int c24; }Params; __kernel __attribute__((reqd_work_group_size(WORK_WIDTH,WORK_HEIGHT,1))) void MyKernel ( __constant Params *params, __global int2 *outb ) { const size_t g1D = get_global_id(1U)*BLOCK_WIDTH + get_global_id(0U); size_t i; for ( i=0; i<ITERATIONS; ++i ) { const int val = params->c1.x + params->c21 + params->c22 + params->c23 + params->c24; int2 i2v; i2v.x = val; i2v.y = val; outb[g1D+i*BLOCK_SIZE] = i2v; outb[g1D+i*BLOCK_SIZE] += (int2)(1); } }
