From: Francisco Jerez <currojerez@riseup.net>
Date: Sat, 12 May 2012 17:24:09 +0000 (+0200)
Subject: clover: Check the total work-group size provided to clEnqueueNDRangeKernel.
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=fcab4d4a34a28ec153612e41d5850884c5ccbfb9;p=mesa.git

clover: Check the total work-group size provided to clEnqueueNDRangeKernel.
---

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<size_t>(), 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<size_t>(), 1u,
+                  block_size, block_size + dims) >
+             q->dev.max_threads_per_block())
+            throw error(CL_INVALID_WORK_GROUP_SIZE);
+      }
    }
 
    ///