Hello
I have a problem witing to a char array, when using the GPU.
See my code attached.
I use the following:
Ubuntu 9.10
Fglrx 10.1/10.2
AMD stream SDK 2.01.
My output picture is somewhat distorted. See my attached picture. On a nvidia gfx card and using the cpu, there is no problem "copying" the uchar image to another.
I use the correct extension, to be able to write to an uchar pointer array.
#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable __kernel void hello(__global uchar * input, __global uchar * output, __global int * widthstep) { int tidx = get_global_id(0); int tidy = get_global_id(1); output[tidx + tidy*widthstep[0]] = input[tidx + tidy*widthstep[0]]; }
cl_khr_byte_addressable_store is not supported at the moment. but you can try set enviroment variable GPU_BYTE_ADDRESSABLE_STORE=1
Thx for the quick response.
well running:
export = GPU_BYTE_ADDRESSABLE_STORE=1
./opencltest
Doesn't work. It does exactly the same.
no thats wrong.
export GPU_BYTE_ADDRESSABLE_STORE=1
Ups. Well that was what I did. 🙂
It doesn't work unfortunately.
it is undocumented feature. so it no big suprise it didnt work. but most likely it will work only on 5xxx series card. AMD for 4xxx series card do not plan any extension.
Thx for your help.
I will try convert my frame to floats instead.
Hmm I am trying to do a uchar to float typecast. See the code.
I get a lot of corruption in the output image again.
Should I just stay completely away from uchar and opencl?
#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable __kernel void hello(__global uchar * input, __global float * output, __global int * widthstep) { int tidx = get_global_id(0); int tidy = get_global_id(1); output[tidx + tidy*widthstep[0]] = (float)input[tidx + tidy*widthstep[0]]/255.0f; }
Thx Micah.
Are you going to support the cl_khr_byte_addressable_store part, or is it just the uchar to float typecast?
If cl_khr_byte_addressable_store is going to be supported, will the 4xxx series work with that extension?
I tested writing into char pointer and it worked in HD5870.
I can't write into char or short arrays. Neither private nor local memory.
System:
ATI 5770, AMD X2
Catalyst 10.5
SDK 2.1
Ubuntu 9.04 x86_64
smatovic,
Could you provide a testcase?
inside the kernel, integer works:
int board[82];
board[0] = 111;
but char not
char board[82];
board[0] = 111;
clBuildProgram fails.
Please check the buildlog to get the exact error message(you can also use SKA) and post it.
clBuildProgram returns -11.
Is there another way to get an error message?
I thought SKA is only available on Windows.
./my_kernels.cl(108): error: write to < 32 bits via pointer not allowed
unless cl_khr_byte_addressable_store is enabled
board[0] = 111;
^
ok, i think i got it:
#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable
char pointers are working.
clGetProgramBuildInfo can be used to return the verbose build log.