OpenCL variable assignment fails when using arithmetic expressions with variable

Question

**Hello, I´m trying to code a kernel using opencl. But Im stuck with a stranger error inside of the kernel function. The flow is more or less like this:

__kernel function1() {
      struct mytype;
      function2(&mytype);
 }

 function2(struct *mytype) {
      uchar *ptr = mytype->value2;
      function3(ptr);
 }

 function3(byte* ptr) {
      uint16 v1 = 10;
      uint16* ptr2 = (uint16*) ptr;
      *ptr2 = v1 >> 8; 
 }


 struct mytype {
       uchar value1[8];
       uchar value2[8];
       uint key[52];
       uint bufleft;
  }

The code fails when executing the assignment:

*ptr2 = v1 >> 8;

But the only message I receive is "clFlush(): CL_UNKNOWN_ERROR" If I try to assign a value and not an expression, than it works. I´m using OpenCL 1.2 CUDA in Ubuntu

score 2 · Accepted Answer · answered Jun 19 '18 at 11:04

I think this might be undefined behaviour; if you need to reference the same memory as 2 different types, use a union type. Note that uint16 is a vector of 16 uints, not a ushort (16-bit unsigned integer), and needs to be aligned accordingly. value2 is only guaranteed to be aligned on a 4-byte boundary (because of the uint members of the struct) so this won't be enough.

score 0 · Answer 2 · answered Jun 18 '18 at 22:12

I am running OpenCL 1.0 on Intel's FPGA SDK for OpenCL (emulator). I had to tweak your code for it to compile, I hopefully didn't change something essential.

`struct mytype {
   uchar value1[8];
   uchar value2[8];
   uint key[52];
   uint bufleft;
}; 
void function3(char* ptr) { //I used uchar * ptr too just for kicks, same result
  uint16 v1 = 10;
  uint16* ptr2 = (uint16*) ptr;
  *ptr2 = v1 >> 8; 
}
void function2(struct mytype * a) {
  uchar *ptr = a->value2;
  function3(ptr);
}
__kernel void function1() {
  struct mytype b;
  function2(&b);
}

It ran smooth, with no runtime errors, even when I made sure to use -O0. Because of this, I cannot find the exact solution, but after perusing the OpenCL 1.2 Specification, I believe your issue may have to do with Alignment. On page 239 you can find alignment attributes for structs.

I will edit this answer as I dig deeper into the matter, so stay tuned and feel free to edit.

OpenCL variable assignment fails when using arithmetic expressions with variable

2 Answers2