Crash under particular circumstances

Discussion created by kazocsaba on Feb 9, 2010
Latest reply on Feb 12, 2010 by genaganna

I've spent quite some time stripping down my code to produce the simplest test case for a crash I'm encountering. Hoping I can help.

The OpenCL code is attached. Run it with a float buffer of size 1, global_work_size=[80,80,80], local_work_size=[4,4,4].

There's quite a list of changes that can be made that (individually, independently from one another) stop the crash from occuring:

  • reducing the global work size in any dimension by 4
  • reducing the local work size in any dimension by 2
  • declaring an array of less than 9 elements
  • changing the array type to float
  • removing the first write to the array or changing the order of the two writes (which particular indices are written is irrelevant)
  • replacing the function parameter f[0] by a constant expression
  • using abs(0) instead of fabs(0)
  • replacing -x in the ?: construct by x or 1*x (x+1, x/1 crash)
  • declaring a variable to hold the value of this expression of x and using this variable in the conditional operator


float thing(float x) { return (fabs(0)<0) ? 0 : -x; } __kernel void test(__global float* f) { float4 array[9]; array[0]=0; array[1]=thing(f[0]); }