Weird result with unsigned char type

OpenCL
1

Hello.

I wrote a very simple kernel that just copies the data from a source to a destination (e.g. image copy).


__kernel void copy_img(
__global unsigned char *src, 
__global unsigned char *dst, 
const int w
) 
{
	int gid_x = get_global_id(0) ;
	int gid_y = get_global_id(1) ;
	int index = gid_x + gid_y * w ;
	
	dst[index] = src[index] ;
}

When I run this kernel with CPU (CL_DEVICE_CPU), it works well. However, the kernel above results in a weird result when it is run on GPU (CL_DEVICE_GPU). Most of the destination array is filled with garbage data…

An interesting thing is that the kernel work well when I changed the data type of destination array from ‘unsigned char *’ to ‘float *’ !!


__kernel void copy_img(
__global unsigned char *src, 
__global float *dst, 
const int w
) 
{
	int gid_x = get_global_id(0) ;
	int gid_y = get_global_id(1) ;
	int index = gid_x + gid_y * w ;
	
	dst[index] = src[index] ;
}

This is very weird…
Is there anything I have to keep in my mind to work with unsigned char type ??

Thanks in advance.

2

What datatype are you actually working with? Are you working on an array of unsigned char or an array of float?

When you use char, each work-item only copies one byte at a time (sizeof(cl_uchar)==1). When using float, each work-item copies four bytes at a time (sizeof(cl_float)==4).

3

What datatype are you actually working with? Are you working on an array of unsigned char or an array of float?

Actually, my data is in unsigned char, a gray image with 8 bits per pixel. So my initial kernel function receives unsigned char array as arguments. When I changed the data type to float* in the kernel, I also changed the image data to a float array.

So there is no confusion in data type…

4

That could a bug in the OpenCL implementation you are using. Since it’s so easy to reproduce, why not reporting this bug to the vendor?