Mesa (main): agx: List sr enum in Python

[GitLab Mirror]

Module: Mesa
Branch: main
Commit: 307b8f1b2f20dcc323f9ed7fa9ae009c355d7fad
URL:    http://cgit.freedesktop.org/mesa/mesa/commit/?id=307b8f1b2f20dcc323f9ed7fa9ae009c355d7fad

Author: Alyssa Rosenzweig <alyssa at rosenzweig.io>
Date:   Sat Jun  5 14:47:53 2021 -0400

agx: List sr enum in Python

Signed-off-by: Alyssa Rosenzweig <alyssa at rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11199>

---

 src/asahi/compiler/agx_compiler.h |  5 -----
 src/asahi/compiler/agx_opcodes.py | 24 +++++++++++++++++++++++-
 2 files changed, 23 insertions(+), 6 deletions(-)

diff --git a/src/asahi/compiler/agx_compiler.h b/src/asahi/compiler/agx_compiler.h
index eb8de8b245f..dc3d6ff2bcf 100644
--- a/src/asahi/compiler/agx_compiler.h
+++ b/src/asahi/compiler/agx_compiler.h
@@ -177,11 +177,6 @@ agx_is_equiv(agx_index left, agx_index right)
 #define AGX_MAX_DESTS 1
 #define AGX_MAX_SRCS 5
 
-enum agx_sr {
-   AGX_SR_INVOCATION,
-   /* stub */
-};
-
 enum agx_icond {
    AGX_ICOND_UEQ = 0,
    AGX_ICOND_ULT = 1,
diff --git a/src/asahi/compiler/agx_opcodes.py b/src/asahi/compiler/agx_opcodes.py
index ba75bd5ec7c..63f7a31222d 100644
--- a/src/asahi/compiler/agx_opcodes.py
+++ b/src/asahi/compiler/agx_opcodes.py
@@ -95,7 +95,29 @@ NEST = immediate("nest")
 INVERT_COND = immediate("invert_cond")
 NEST = immediate("nest")
 TARGET = immediate("target", "agx_block *")
-SR = immediate("sr", "enum agx_sr")
+SR = enum("sr", {
+   0:  'threadgroup_position_in_grid.x',
+   1:  'threadgroup_position_in_grid.y',
+   2:  'threadgroup_position_in_grid.z',
+   4:  'threads_per_threadgroup.x',
+   5:  'threads_per_threadgroup.y',
+   6:  'threads_per_threadgroup.z',
+   8:  'dispatch_threads_per_threadgroup.x',
+   9:  'dispatch_threads_per_threadgroup.y',
+   10: 'dispatch_threads_per_threadgroup.z',
+   48: 'thread_position_in_threadgroup.x',
+   49: 'thread_position_in_threadgroup.y',
+   50: 'thread_position_in_threadgroup.z',
+   51: 'thread_index_in_threadgroup',
+   52: 'thread_index_in_subgroup',
+   53: 'subgroup_index_in_threadgroup',
+   56: 'active_thread_index_in_quad',
+   58: 'active_thread_index_in_subgroup',
+   62: 'backfacing',
+   80: 'thread_position_in_grid.x',
+   81: 'thread_position_in_grid.y',
+   82: 'thread_position_in_grid.z',
+})
 
 FUNOP = lambda x: (x << 28)
 FUNOP_MASK = FUNOP((1 << 14) - 1)