OpenCL を使用して LBM を並列化していますが、カーネルへの引数としてカスタム データ型のカーネル ヘッダー ファイルがどのように生成されるかに関して問題があります。必要に応じてカーネル ファイル ( rebound.cl
)内でデータ型を定義します ( ) typedef struct {...} t_speed;
。データ型t_speed
はヘッダー ファイルで生成されますが、これは明らかに構文的に正しくなく、その後ビルドが失敗します。これは重大な問題というより厄介な問題ですが、修正することで多くの時間を節約できます!
カーネルファイル: 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 */
}
}
}
カーネル ヘッダー ファイル: 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);