This kernel runs fine on my computer on a CPU device and the CodeXL Kernel Analyzer verifies it should work on a wide variety of cards - everything HD5k or newer for image support. I'm unfortunately currently stuck with a HD4870 and therefore cannot test it myself. On my roommates computer, however, clBuildProgram reports an unexpected exception, both in my own host application and the Kernel Analyzer. We installed new drivers, the APP SDK and CodeXL but nothing seems to work. Is this a driver bug or is the code actually somehow wrong?
constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;
typedef struct {
float4 orig;
float4 dir;
} CLRay;
typedef struct {
float2 uv;
float t;
int id;
} CLResult;
typedef struct {
float4 a1, a2, a3;
} CLTriangle;
typedef struct {
float4 nor, col;
} CLTriangledata;
kernel void fill_image( image2d_t img,
constant CLRay * ray,
constant CLResult * result,
global CLTriangledata * triangledata,
float N) {
const int2 id = (int2){get_global_id(0), get_global_id(1)};
const int id1d = id.y*get_global_size(0)+id.x;
int res = result[id1d].id;
float4 col = (float4).0f;
if(res>=0)
col = pow(triangledata[res].col, 2.2f) * fabs(dot(ray[id1d].dir, triangledata[res].nor));
float4 oldcol = read_imagef(img, sampler, id);
write_imagef(img, id, (oldcol*(N-1.0f)+col)/N);
}
