From fcab4d4a34a28ec153612e41d5850884c5ccbfb9 Mon Sep 17 00:00:00 2001 From: Francisco Jerez Date: Sat, 12 May 2012 19:24:09 +0200 Subject: [PATCH] clover: Check the total work-group size provided to clEnqueueNDRangeKernel. --- .../state_trackers/clover/api/kernel.cpp | 27 ++++++++++++------- 1 file changed, 17 insertions(+), 10 deletions(-) diff --git a/src/gallium/state_trackers/clover/api/kernel.cpp b/src/gallium/state_trackers/clover/api/kernel.cpp index 44eeb277127..ab4982d2951 100644 --- a/src/gallium/state_trackers/clover/api/kernel.cpp +++ b/src/gallium/state_trackers/clover/api/kernel.cpp @@ -223,16 +223,23 @@ namespace { if (!grid_size || any_of(is_zero(), grid_size, grid_size + dims)) throw error(CL_INVALID_GLOBAL_WORK_SIZE); - if (block_size && any_of([](size_t b, size_t max) { - return b == 0 || b > max; - }, block_size, block_size + dims, - q->dev.max_block_size().begin())) - throw error(CL_INVALID_WORK_ITEM_SIZE); - - if (block_size && any_of([](size_t b, size_t g) { - return g % b; - }, block_size, block_size + dims, grid_size)) - throw error(CL_INVALID_WORK_GROUP_SIZE); + if (block_size) { + if (any_of([](size_t b, size_t max) { + return b == 0 || b > max; + }, block_size, block_size + dims, + q->dev.max_block_size().begin())) + throw error(CL_INVALID_WORK_ITEM_SIZE); + + if (any_of([](size_t b, size_t g) { + return g % b; + }, block_size, block_size + dims, grid_size)) + throw error(CL_INVALID_WORK_GROUP_SIZE); + + if (fold(std::multiplies(), 1u, + block_size, block_size + dims) > + q->dev.max_threads_per_block()) + throw error(CL_INVALID_WORK_GROUP_SIZE); + } } /// -- 2.30.2