OpenCL & Xcode - Incorrect kernel header being gen

2019-08-10 21:49发布

I'm parallelising a LBM using OpenCL and have across a problem regarding how the kernel header files are being generated for a custom data type as an argument to the kernel. I define the data type within the kernel file (rebound.cl) as required (typedef struct {...} t_speed;) and the data type t_speed is generated in the header file which is obviously syntactically incorrect and the build subsequently fails. Whilst this is more of an annoyance than a major problem, fixing it would save a lot of time!

Kernel file: rebound.cl

#ifdef cl_khr_fp64
#pragma OPENCL EXTENSION cl_khr_fp64 : enable
#elif defined(cl_amd_fp64)
#pragma OPENCL EXTENSION cl_amd_fp64 : enable
#else
#error "Double precision floating point not supported by OpenCL implementation."
#endif

#define NSPEEDS 5
typedef struct {
    double speeds[NSPEEDS];
} t_speed;

__kernel void rebound (__global t_speed* cells,
                       __global t_speed* tmp_cells,
                       __global const unsigned char* obstacles,
                       const unsigned short int count)
{
    int i = get_global_id(0);
    if (i < count) {
        if (obstacles[i]) {
            cells[i].speeds[1] = tmp_cells[i].speeds[3]; /* East -> West */
            cells[i].speeds[3] = tmp_cells[i].speeds[1]; /* West -> East*/
            cells[i].speeds[2] = tmp_cells[i].speeds[4]; /* North -> South */
            cells[i].speeds[4] = tmp_cells[i].speeds[2]; /* South -> North */
        }
    }
}

Kernel header file: rebound.cl.h

/***** GCL Generated File *********************/
/* Automatically generated file, do not edit! */
/**********************************************/

#include <OpenCL/opencl.h>
typedef struct  {
  double [5] speeds;
} _t_speed_unalign;

typedef _t_speed_unalign __attribute__ ((aligned(8))) t_speed;

extern void (^rebound_kernel)(const cl_ndrange *ndrange, t_speed* cells, t_speed* tmp_cells, cl_uchar* obstacles, cl_ushort count);

0条回答
登录 后发表回答