I always get return code -51 when I try to set the second argument of my kernel. This argument is a local buffer. Here is the call I make: clSetKernelArg(in_Kernel.mKernel, 1, 1024, NULL); The kernel prototype is: __kernel void GetEfficiency( __constant const SimulationParams * const in_pSimulationParams, __local float * l_afEfficiencies, __global const float * const in_pRandomValues, __global float * const out_pfEfficiencies) So argument index 1 is really a local buffer as can be seen from the prototype. I will try to investigate this further.
Created attachment 82462 [details] Addess spaces support for kernel arguments Use the data from getAddressSpaceMap to determine the correct type of the kernel arguments.
(In reply to comment #1) > Created attachment 82462 [details] > Addess spaces support for kernel arguments > > Use the data from getAddressSpaceMap to determine the correct type of the > kernel arguments. This patch looks good to me, can you send it to mesa-dev@lists.freedesktop.org. I think for r600g to work correctly, you will also need to make sure we are writing the correct lds size to the SQ_LDS_ALLOC register. I think what we want is the size reported by clover plus the size reported by the compiler.
Created attachment 82485 [details] [review] Addess spaces support for kernel arguments Added the local argument size to lds alloc size.
Created attachment 82498 [details] [review] Addess spaces support for kernel arguments Fixed the local size to be in dw
(In reply to comment #4) > Created attachment 82498 [details] [review] [review] > Addess spaces support for kernel arguments > > Fixed the local size to be in dw This patch looks good to me, can you send all the paches to mesa-dev@lists.freedesktop.org for review?
This was fixed in git a while ago. commit 4f8048bb5a8558ae4313b12ffd70b593cc629fe8 commit d9576598c7e1c6e4fee913a918345190248a9d19
Use of freedesktop.org services, including Bugzilla, is subject to our Code of Conduct. How we collect and use information is described in our Privacy Policy.