OpenCLKernel1¶
-
MLModule¶ author Felix Ritterpackage FMEwork/ReleaseMeVisdll MLOpenCLKernel1definition MLOpenCLKernel1.def see also Convolutionkeywords OpenCL,GLSL,convolution,kernel,arithmetic
Purpose¶
The OpenCLKernel1 module applies an OpenCL kernel to an ML image and may be used to process the image data on OpenCL-supported devices.
Usage¶
Connect the module to the image pipeline and type an OpenCL kernel into the kernel text field.
Details¶
The kernel below examplifies the integration of an image processing algorithm using OpenCL:
#define KERNEL_SIZE 9
__kernel void convolve (
const __global float* input,
__global float* output,
const uint8 img_size)
{
__local float cKernel[KERNEL_SIZE];
__local int offset[KERNEL_SIZE];
cKernel[0] = 0.0; cKernel[1] = 1.0; cKernel[2] = 0.0;
cKernel[3] = 1.0; cKernel[4] = -4.0; cKernel[5] = 1.0;
cKernel[6] = 0.0; cKernel[7] = 1.0; cKernel[8] = 0.0;
offset[0] = -img_size.s0 - 1;
offset[1] = -img_size.s0 + 0;
offset[2] = -img_size.s0 + 1;
offset[3] = -1;
offset[4] = +0;
offset[5] = +1;
offset[6] = img_size.s0 - 1;
offset[7] = img_size.s0 + 0;
offset[8] = img_size.s0 + 1;
const int i = get_global_id(0);
unsigned int o;
float sum = 0;
for (int ki = 0; ki < KERNEL_SIZE; ++ki) {
o = i + offset[ki];
if (o > 0 && o < img_size.s6) {
sum += input[o] * cKernel[ki];
}
}
output[i] = sum;
}
The name of the function is of no particular importance, the module uses the first __kernel function it discovers in the code. Input and output pointers address the input and output buffers which store single precision floating point data. The third, uint8 parameter stores eight single integer parameters with the following meaning:
- .s0 … .s5
- the six output image dimensions as provided by the ML
- .s6
- the total number of voxels in the output image
- .s7
- .s7 := .s0*.s1*.s2 to conveniently access the tree dimensional volume