From: Jason Ekstrand <jason.ekstrand@intel.com>
Date: Wed, 1 Jun 2016 18:20:22 +0000 (-0700)
Subject: nir/spirv: Handle the WorkgroupSize builtin decoration
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=1f7b54ed299bea95f774e7d8baa181c11118b3fe;p=mesa.git

nir/spirv: Handle the WorkgroupSize builtin decoration

This fixes the 7 dEQP-VK.pipeline.spec_constant.compute.local_size.* tests
in the latest dev version of the Vulkan CTS.

Signed-off-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Cc: "12.0" <mesa-stable@lists.freedesktop.org>
---

diff --git a/src/compiler/spirv/spirv_to_nir.c b/src/compiler/spirv/spirv_to_nir.c
index bb7aba4d9d4..cece645c6bc 100644
--- a/src/compiler/spirv/spirv_to_nir.c
+++ b/src/compiler/spirv/spirv_to_nir.c
@@ -931,6 +931,25 @@ get_specialization(struct vtn_builder *b, struct vtn_value *val,
    return const_value;
 }
 
+static void
+handle_workgroup_size_decoration_cb(struct vtn_builder *b,
+                                    struct vtn_value *val,
+                                    int member,
+                                    const struct vtn_decoration *dec,
+                                    void *data)
+{
+   assert(member == -1);
+   if (dec->decoration != SpvDecorationBuiltIn ||
+       dec->literals[0] != SpvBuiltInWorkgroupSize)
+      return;
+
+   assert(val->const_type == glsl_vector_type(GLSL_TYPE_UINT, 3));
+
+   b->shader->info.cs.local_size[0] = val->constant->value.u[0];
+   b->shader->info.cs.local_size[1] = val->constant->value.u[1];
+   b->shader->info.cs.local_size[2] = val->constant->value.u[2];
+}
+
 static void
 vtn_handle_constant(struct vtn_builder *b, SpvOp opcode,
                     const uint32_t *w, unsigned count)
@@ -1151,6 +1170,9 @@ vtn_handle_constant(struct vtn_builder *b, SpvOp opcode,
    default:
       unreachable("Unhandled opcode");
    }
+
+   /* Now that we have the value, update the workgroup size if needed */
+   vtn_foreach_decoration(b, val, handle_workgroup_size_decoration_cb, NULL);
 }
 
 static void