Cannot use custom type in kernel

[GRodr25]

Hello,

I am unable to use a custom type in my OpenCL kernel, passing it by value (error: unsupported kernel argument type. This problem arises in compilation time). I found a solution in this question https://forums.intel.com/s/question/0D50P00004Nr1jn/emulation-fails-with-userdefined-data-type-for-a-kernel-argument, which tackles the problem passing a pointer and setting the address qualifier __global. I would rather stay with the passing by value version, due to some application constraints. I am using Intel FPGA SDK for OpenCL 16.1. Is there other way to overcome this issue instead of this ugly trick?

Just in case it helps you, I reproduced the problem with a minimal kernel:

typedef struct {   char a;   char b;  } error_type;    __kernel void min_kernel(error_type err_data) {   err_data.a = 'a';   err_data.b = 'b'; }

Thank you.

[MEIYAN_L_Intel]

Hi,

I had compare two OpenCL SDK version which are v16.0 and v19.3.

The latest version had stated "Pass structure parameters (struct) in OpenCL kernels either by value or as a pointer to a structure" while in v16.0 stated "Convert each structure parameter (struct) to a pointer that points to a structure".

I am checking internally whether in v16.0 can passing structure parameters by value.

Thanks

[MEIYAN_L_Intel]

Hi,

You can be able to put the struct as a parameter in the kernel declaration and then call clSetKernelArg() from host (https://www.khronos.org/registry/OpenCL/sdk/1.2/docs/man/xhtml/clSetKernelArg.html).

For an example:

// Host

error_type et;

clSetKernelArg( min_kernel, 0, sizeof(error_type), &(et) )

Thanks

[GRodr25]

Hello @MeiYanL_Intel​,

Thank your for your answer. I am afraid that will not work, since AOC will not allow me to compile min_kernel. I read in one of the Intel guides structs must be passed as pointers, indeed. So, there seems to be no other way than creating a buffer for the struct and passing a pointer to it.

Thank you.

[MEIYAN_L_Intel]

Hi,

You are welcome.

Thanks