From 9d5b05f28adb433af2c6720ee7a602c9f0581fd3 Mon Sep 17 00:00:00 2001 From: foreman Date: Tue, 9 Jun 2015 12:58:27 -0400 Subject: [PATCH] P4 to Git Change 1159349 by rili@rili_opencl_stg on 2015/06/09 12:52:32 EPR #419313 - Blit Kernel execution split Enabled flag GPU_ENABLE_LARGE_ALLOCATION Return error if global work size is larger than 32bits. Code Review#7664 Affected files ... ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_execute.cpp#22 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpukernel.cpp#285 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpukernel.hpp#111 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuvirtual.cpp#365 edit ... //depot/stg/opencl/drivers/opencl/runtime/utils/flags.hpp#234 edit [ROCm/clr commit: aca7118ed45e9c37e69385991f5a280c5cf7e791] --- projects/clr/opencl/api/opencl/amdocl/cl_execute.cpp | 8 ++++++++ 1 file changed, 8 insertions(+) diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_execute.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_execute.cpp index 95c5c7a85f..42a4511639 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_execute.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_execute.cpp @@ -205,6 +205,14 @@ RUNTIME_ENTRY(cl_int, clEnqueueNDRangeKernel, ( if (global_work_size == NULL) { return CL_INVALID_VALUE; } + else { + // >32bits global work size is not supported. + for (cl_uint dim = 0; dim < work_dim; ++dim) { + if (global_work_size[dim] > static_cast(0xffffffff)) { + return CL_INVALID_GLOBAL_WORK_SIZE; + } + } + } if (local_work_size == NULL) { static size_t zeroes[3] = { 0, 0, 0 };