fdo-mirrors/mesa
1
0
Fork 0
mirror of https://gitlab.freedesktop.org/mesa/mesa.git synced 2026-05-16 01:08:05 +02:00
Code Issues Projects Releases Packages Wiki Activity Actions 1

rusticl/kernel: return CL_INVALID_WORK_GROUP_SIZE in clEnqueueNDRangeKernel for an explicit 0 workgroup

Browse source 
See https://github.com/KhronosGroup/OpenCL-Docs/pull/1542

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/41358>
This commit is contained in:
Karol Herbst 2026-05-05 18:04:51 +02:00 • committed by Marge Bot
parent f69fe4cd55
commit ac5db2d42e
1 changed files with 8 additions and 2 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

10
src/gallium/frontends/rusticl/api/kernel.rs
View file

@ -699,7 +699,7 @@ fn enqueue_ndrange_kernel(
work_dim: cl_uint,
global_work_offset: *const usize,
global_work_size: *const usize,
local_work_size: *const usize,
local_work_size_ptr: *const usize,
num_events_in_wait_list: cl_uint,
event_wait_list: *const cl_event,
event: *mut cl_event,
@ -732,9 +732,15 @@ fn enqueue_ndrange_kernel(
// we assume the application gets it right and doesn't pass shorter arrays then actually needed.
let global_work_size = unsafe { kernel_work_arr_or_default(global_work_size, work_dim) };
let local_work_size = unsafe { kernel_work_arr_or_default(local_work_size, work_dim) };
let local_work_size = unsafe { kernel_work_arr_or_default(local_work_size_ptr, work_dim) };
let global_work_offset = unsafe { kernel_work_arr_or_default(global_work_offset, work_dim) };
// CL_INVALID_WORK_GROUP_SIZE if local_work_size is not NULL and if the total number of
// work-items in the work-group is zero
if !local_work_size_ptr.is_null() && local_work_size.iter().any(|&l| l == 0) {
return Err(CL_INVALID_WORK_GROUP_SIZE);
}
let device_bits = q.device.address_bits();
let device_max = u64::MAX >> (u64::BITS - device_bits);