tgsi: Add WORK_DIM System Value

Add a new WORK_DIM SV type, this is will return the grid dimensions
(1-4) for compute (opencl) kernels.

This is necessary to implement the opencl get_work_dim() function.

Signed-off-by: Hans de Goede <hdegoede@redhat.com>
Reviewed-by: Ilia Mirkin <imirkin@alum.mit.edu>
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>

3 files changed, 10 insertions, 0 deletions

diff --git a/src/gallium/auxiliary/tgsi/tgsi_strings.c b/src/gallium/auxiliary/tgsi/tgsi_strings.c
index c13f7eaeccaa..536a4c8f3a93 100644
--- a/src/gallium/auxiliary/tgsi/tgsi_strings.c
+++ b/src/gallium/auxiliary/tgsi/tgsi_strings.c
@@ -99,6 +99,7 @@ const char *tgsi_semantic_names[TGSI_SEMANTIC_COUNT] =
    "HELPER_INVOCATION",
    "BASEINSTANCE",
    "DRAWID",
+   "WORK_DIM",
 };
 
 const char *tgsi_texture_names[TGSI_TEXTURE_COUNT] =
diff --git a/src/gallium/docs/source/tgsi.rst b/src/gallium/docs/source/tgsi.rst
index bf094068e881..c6e5ceb7eceb 100644
--- a/src/gallium/docs/source/tgsi.rst
+++ b/src/gallium/docs/source/tgsi.rst
@@ -3046,6 +3046,14 @@ For vertex shaders, the zero-based index of the current draw in a
 component is used.
 
 
+TGSI_SEMANTIC_WORK_DIM
+""""""""""""""""""""""
+
+For compute shaders started via opencl this retrieves the work_dim
+parameter to the clEnqueueNDRangeKernel call with which the shader
+was started.
+
+
 Declaration Interpolate
 ^^^^^^^^^^^^^^^^^^^^^^^
 
diff --git a/src/gallium/include/pipe/p_shader_tokens.h b/src/gallium/include/pipe/p_shader_tokens.h
index abf26d3b4683..7621ab92fac9 100644
--- a/src/gallium/include/pipe/p_shader_tokens.h
+++ b/src/gallium/include/pipe/p_shader_tokens.h
@@ -199,6 +199,7 @@ enum tgsi_semantic {
    TGSI_SEMANTIC_HELPER_INVOCATION,  /**< current invocation is helper */
    TGSI_SEMANTIC_BASEINSTANCE,
    TGSI_SEMANTIC_DRAWID,
+   TGSI_SEMANTIC_WORK_DIM,    /**< opencl get_work_dim value */
    TGSI_SEMANTIC_COUNT,       /**< number of semantic values */
 };