Reputation: 3670
I have sample code but it completely leaves out what my (void*)should_be!
I setup a cl_image_desc, cl_image_format, buffer, origin, and region:
cl_image_desc desc;
desc.image_type = CL_MEM_OBJECT_IMAGE2D;
desc.image_width = width;
desc.image_height = height;
desc.image_depth = 0;
desc.image_array_size = 0;
desc.image_row_pitch = 0;
desc.image_slice_pitch = 0;
desc.num_mip_levels = 0;
desc.num_samples = 0;
desc.buffer = NULL;
cl_image_format format;
format.image_channel_order = CL_R;
format.image_channel_data_type = CL_FLOAT;
cl_mem bufferSourceImage = clCreateImage(context, CL_MEM_READ_ONLY, &format, &desc, NULL, NULL);
size_t origin[3] = {0, 0, 0};
size_t region[3] = {width, height,1};
In this next snippet sourceImage is a void pointer to my image. But what is my image? For every pixel there are r, g, b, a, x, and y values.
clEnqueueWriteImage(queue, bufferSourceImage, CL_TRUE, origin, region, 0, 0, sourceImage, 0, NULL, NULL);
How do I turn my image (a bunch of (r,g,b,a,x,y)'s) into a suitable array?
This is the kernel they provide:
__kernel void convolution(__read_only image2d_t sourceImage, __write_only image2d_t outputImage, int rows, int cols, __constant float* filter, int filterWidth, sampler_t sampler)
{
int column = get_global_id(0);
int row = get_global_id(1);
int halfWidth = (int)(filterWidth/2);
float4 sum = {0.0f, 0.0f, 0.0f, 0.0f};
int filterIdx = 0;
int2 coords;
for(int i = -halfWidth; i <= halfWidth; i++)
{
coords.y = row + i;
for(int i2 = -halfWidth; i2 <= halfWidth; i2++)
{
coords.x = column + i2;
float4 pixel;
pixel = read_imagef(sourceImage, sampler, coords);
sum.x += pixel.x * filter[filterIdx++];
}
}
if(myRow < rows && myCol < cols)
{
coords.x = column;
coords.y = row;
write_imagef(outputImage, coords, sum);
}
}
Upvotes: 1
Views: 2221
Reputation: 10719
Set up the cl_image_format as you like and then you just have to follow that format what you selected. Currently your channel (R, G, B, A) data should be represented as "single precision floating-point value" - image_channel_data_type = CL_FLOAT
, and you can take only one channel of those and feed it into the expected R channel (image_channel_order = CL_R
).
Your kernel expect float:
float4 pixel;
pixel = read_imagef(sourceImage, sampler, coords);
Upvotes: 1