asahi: handle load_workgroup_size

Signed-off-by: Karol Herbst <git@karolherbst.de>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25052>

diff --git a/src/asahi/compiler/agx_compile.c b/src/asahi/compiler/agx_compile.c
index d4fa491..26799c5 100644 (file)
--- a/src/asahi/compiler/agx_compile.c
+++ b/src/asahi/compiler/agx_compile.c
@@ -1077,6 +1077,10 @@ agx_emit_intrinsic(agx_builder *b, nir_intrinsic_instr *instr)
       return agx_load_compute_dimension(b, dst, instr,
                                         AGX_SR_THREADGROUP_POSITION_IN_GRID_X);
 
+   case nir_intrinsic_load_workgroup_size:
+      return agx_load_compute_dimension(b, dst, instr,
+                                        AGX_SR_THREADS_PER_THREADGROUP_X);
+
    case nir_intrinsic_load_global_invocation_id:
       return agx_load_compute_dimension(b, dst, instr,
                                         AGX_SR_THREAD_POSITION_IN_GRID_X);