Commit graph

1620 commits

Author SHA1 Message Date
Konstantin Seurer
b962063d72 nir: Remove nir_parallel_copy_instr
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36483>
2025-11-04 18:51:51 +00:00
Marek Olšák
2f6b4803ab nir/validate: expand IO intrinsic validation with nir_io_semantics
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
There are many workarounds.

v2: add more validation

Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com> (v1)
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/38113>
2025-11-02 02:21:46 +00:00
Daniel Schürmann
f61cd64af8 nir/builder: add option to immediately constant-fold ALU instructions upon insertion
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37195>
2025-10-30 19:28:07 +00:00
Rhys Perry
92beca9aa5 nir/lower_tex: optimize txd(coord, ddx/ddy(coord))
fossil-db (gfx1201):
Totals from 73 (0.09% of 79839) affected shaders:
MaxWaves: 1668 -> 1670 (+0.12%)
Instrs: 352537 -> 347991 (-1.29%); split: -1.29%, +0.00%
CodeSize: 1924140 -> 1887660 (-1.90%); split: -1.90%, +0.00%
VGPRs: 6360 -> 6324 (-0.57%)
Latency: 3891330 -> 3888192 (-0.08%); split: -0.10%, +0.02%
InvThroughput: 789998 -> 783583 (-0.81%); split: -0.84%, +0.03%
VClause: 6409 -> 6408 (-0.02%); split: -0.06%, +0.05%
SClause: 4071 -> 4102 (+0.76%); split: -0.10%, +0.86%
Copies: 16756 -> 16316 (-2.63%); split: -2.94%, +0.32%
PreVGPRs: 5456 -> 5432 (-0.44%); split: -0.57%, +0.13%
VALU: 232982 -> 228117 (-2.09%)
SALU: 32853 -> 32848 (-0.02%); split: -0.05%, +0.03%
VMEM: 9234 -> 9237 (+0.03%)

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37561>
2025-10-23 11:21:59 +00:00
Simon Perretta
ff51e6dc9e nir: commonize barycentric intrinsic opt pass
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
Introduces an opt pass that attempts to optimize
load_barycentric_at_{sample,offset} with simpler load_barycentric_*
equivalents where possible, and optionally lowers
load_barycentric_at_sample to load_barycentric_at_offset with a position
derived from the sample ID instead.

Signed-off-by: Simon Perretta <simon.perretta@imgtec.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37658>
2025-10-22 16:48:01 +00:00
Aitor Camacho
f711c3afed nir: Add KosmicKrisp required utilities
Reviewed-by: Alyssa Anne Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37520>
2025-10-20 16:22:00 +00:00
Emma Anholt
d01aae2fb1 nir: Add a shader bisect tool.
When you're trying to figure out what shader some NIR pass broke, use
nir_shader_bisect_select() to decide between NIR pass behaviors, and then
nir_shader_bisect.py will help you automatically bisect down to which
source_blake3 is at fault.  Once it's identified, it prints you a C call
you can use for selecting that shader specifically, which you can use for
continuing on in your debugging.

On a test I was looking at, this took 10 steps to bisect 134 shaders down
to the source_blake3 of the NIR shader in question.

This idea is heavily lifted from Job Noorman's ir3_shader_bisect.

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37468>
2025-10-09 17:56:30 +00:00
Marek Olšák
3fe651f607 nir: remove load_smem_amd
replaced by load_global_amd + ACCESS_SMEM_AMD

Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36936>
2025-10-08 08:54:11 +00:00
Georg Lehmann
b0d3db3733 nir: add atomic isub
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37702>
2025-10-07 14:07:56 +00:00
Lionel Landwerlin
94f8d0072d nir: add pass to propagate image format to intrinsics
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Alyssa Anne Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36773>
2025-10-07 08:54:26 +00:00
Lionel Landwerlin
0922a0dd50 nir/lower_tex: remove unused options
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Alyssa Anne Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37692>
2025-10-03 20:19:03 +00:00
Lionel Landwerlin
97dde5bc10 nir/lower_tex: add an callback to lower txd ops
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Alyssa Anne Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37692>
2025-10-03 20:19:02 +00:00
Simon Perretta
2a7ebf2ae0 nir/lower_alpha: extend to support dynamic a2c
Signed-off-by: Simon Perretta <simon.perretta@imgtec.com>
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37512>
2025-09-30 12:15:53 +00:00
Simon Perretta
83aecc8f3f mesa/st, nir: commonize unlower_io_to_vars pass
Signed-off-by: Simon Perretta <simon.perretta@imgtec.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37540>
2025-09-27 23:45:54 +01:00
Aleksi Sapon
8949473023 nir: Fix nir.h MSVC compilation for C++ source files
This kind of C initializer is not accepted by MSVC in C++ mode.

Fixed: 75292ae7 ("nir: Fix gnu-empty-initializer warning ")
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37604>
2025-09-26 18:25:22 +00:00
Aleksi Sapon
75292ae7e4 nir: Fix gnu-empty-initializer warning
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
This also causes a build error on older MSVC.

Fixes: 75381670 ("nir,rusticl: NIR_PASS/nir_pass! validation fixes and improvements")
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37569>
2025-09-25 18:14:22 +00:00
Rhys Perry
7538167096 nir: add NIR_DEBUG=progress_validation
Fails if a shader was changed but the pass didn't report progress.

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Marek Olšák <maraeo@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35069>
2025-09-24 08:20:28 +00:00
Rhys Perry
706ba80057 nir: fix NIR_DEBUG=extended_validation
This broke after divergence became metadata because the divergence
analysis pass does not support all instructions.

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Marek Olšák <maraeo@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35069>
2025-09-24 08:20:28 +00:00
Georg Lehmann
714a149396 nir: remove unsigned upper bound config
All config information is now either in nir->info or nir->options.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37361>
2025-09-16 09:24:04 +00:00
Georg Lehmann
bb67dae12d nir/uub: remove max_workgroup_size from config
For most hardware, this is the same as max invocations in the workgroup.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37361>
2025-09-16 09:24:04 +00:00
Georg Lehmann
f3c08c9d27 nir/uub: use shader_info subgroup size
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37361>
2025-09-16 09:24:04 +00:00
Georg Lehmann
112e160946 nir/uub: remove vertex input handling
Unused since radv started lowering vertex inputs in NIR.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37361>
2025-09-16 09:24:04 +00:00
Georg Lehmann
95c2a65662 nir: remove unused shader_info param in nir_create_shader
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37258>
2025-09-12 21:05:17 +00:00
Georg Lehmann
08b58c3fac nir/lower_subgroups: remove lower_fp64 option
This was incorrect (it also lowered int64 reductions/scans), and the only
user can just use the general callback to precisely only lower what it wants.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37164>
2025-09-09 11:09:22 +00:00
Mel Henning
17876a00af nir: Add a faster lowest common ancestor algorithm
On a fossil from the blender 4.5.0 vulkan backend, this improves compile
times in nak by about 17%. Compile time of other shaders improves by a
more modest 1.2%.

No stat changes on shader-db.

Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36184>
2025-09-08 23:03:13 +00:00
Timothy Arceri
8b1d48cf0b nir: move nir_lower_point_size_mov() to st
Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37037>
2025-09-07 23:13:24 +00:00
Timothy Arceri
450419c3f4 nir: move nir_lower_alpha_test() to the st
This is gl specific and a following fix will add more gl specific
params so here we move it to the st to avoid filling nir.h with
more junk.

Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37037>
2025-09-07 23:13:23 +00:00
Timothy Arceri
8417f4a8eb nir: move nir_lower_drawpixels() to the state tracker
This is gl specific and a following fix will add more gl specific
params so here we move it to the st to avoid filling nir.h with
more junk.

Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37037>
2025-09-07 23:13:22 +00:00
Georg Lehmann
2725eaf9a2 nir/lower_subgroups: change filter to intrinsic callback
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37178>
2025-09-04 14:04:00 +00:00
Job Noorman
e78bd88a06 nir/opt_offsets: add callback to set need_nuw per intrinsic
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
Wether need_nuw is used is currently decided in two different ways:
- globally through the allow_offset_wrap option;
- per intrinsic but hard-coded in opt_offsets.

Make this more flexible by creating a callback that is called per
intrinsic. This will allow backends to decide, on a per-intrinsic basis,
whether need_nuw is needed.

Note that the main use case for ir3 is to add support for opt_offsets
for global memory accesses. Other intrinsics don't need need_nuw but
global memory accesses do.

Signed-off-by: Job Noorman <jnoorman@igalia.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37114>
2025-09-01 11:25:07 +00:00
Job Noorman
bc03086320 nir/opt_offsets: rename max_offset_data to cb_data
We want to add more callbacks and pass the same data.

Signed-off-by: Job Noorman <jnoorman@igalia.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37114>
2025-09-01 11:25:07 +00:00
Faith Ekstrand
26e32417b9 nir: Add an option to make lower_phis_to_regs_block() less clever
Right now it tries to place reg_write instructions as far up the
predecessor chain as possible.  This is useful for a bunch of the passes
that call it since it ensures they don't get placed in dead blocks or in
single successors and things like that.  But it screws up NAK's control
flow lowering so we need the option to turn it off and make the pass
place the reg_write instructions in the most obvious place possible.

Fixes: b013d54e4f ("nak/lower_cf: Flag phis as convergent when possible")
Reviewed-by: Mel Henning <mhenning@darkrefraction.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36914>
2025-08-29 01:24:56 +00:00
Dave Airlie
7a96a928a2 nir: add coop mat flexible dimensions lowering.
This adds a generic lowering pass for coop mat flexible dimensions.

This should be suitable for all drivers that implement coop mat2 flexible dimensions
or even just lowering sw exposed sizes to hw sizes.

Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36544>
2025-08-25 18:55:08 +00:00
Konstantin Seurer
9df7b48d2f nir: Use nir_def_as_* in more places
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36746>
2025-08-24 14:03:09 +00:00
Marek Olšák
3aadae22ad nir: make nir_block::predecessors & dom_frontier sets non-malloc'd
We can just place the set structures inside nir_block.

This reduces the number of ralloc calls by 6.7% when compiling Heaven
shaders with radeonsi+ACO using a release build (i.e. not including
nir_validate set allocations, which are also removed).

Reviewed-by: Gert Wollny <gert.wollny@collabora.com>
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36728>
2025-08-21 06:13:48 +00:00
Marek Olšák
81cb571642 nir/dominance: eliminate ralloc overhead for allocating dom_children
This is only 1% of all ralloc calls of Unigine Heaven with the gallium noop
driver, but it's an easy one to get rid of.

Reviewed-by: Gert Wollny <gert.wollny@collabora.com>
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36728>
2025-08-21 06:13:48 +00:00
Job Noorman
355c9b88f7 nir: add some helpers for dealing with offset_shift
For intrinsics supporting offset_shift, dealing with their offset is a
bit tricky as we cannot simply add a byte offset to it anymore (which is
what most passes want to do). This commit adds some helpers to add byte
offsets (and adjusting offset_shift accordingly) so that individual
passes don't have to worry about this.

Signed-off-by: Job Noorman <jnoorman@igalia.com>
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35092>
2025-08-20 07:51:30 +00:00
Job Noorman
ebea9ce825 nir: add nir_src_is_deref helper
Signed-off-by: Job Noorman <jnoorman@igalia.com>
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35092>
2025-08-20 07:51:30 +00:00
Georg Lehmann
8818d7367d nir/opt_load_skip_helpers: optionally handle intrinsics
Some checks are pending
macOS-CI / macOS-CI (dri) (push) Waiting to run
macOS-CI / macOS-CI (xlib) (push) Waiting to run
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36610>
2025-08-12 08:56:37 +00:00
Georg Lehmann
91572a99bb nir: rename to nir_opt_load_skip_helpers and add options struct
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36610>
2025-08-12 08:56:37 +00:00
Alyssa Rosenzweig
1af0897452 nir/lower_subgroups: add lower_fp64 option
This is needed for doubles lowering to do the right thing.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Marek Olšák <maraeo@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36649>
2025-08-08 20:51:03 +00:00
Qiang Yu
196569b1a4 all: rename gl_shader_stage to mesa_shader_stage
It's not only for GL, change to a generic name.

Use command:
  find . -type f -not -path '*/.git/*' -exec sed -i 's/\bgl_shader_stage\b/mesa_shader_stage/g' {} +

Acked-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Acked-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Acked-by: Yonggang Luo <luoyonggang@gmail.com>
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36569>
2025-08-06 10:28:40 +08:00
Marek Olšák
44350bce1f nir: add nir_variable_create_zeroed helper
This will allow us to switch nir_variable from ralloc to gc_ctx,
which uses a slab allocator.

Reviewed-by: Timothy Arceri <tarceri@itsqueeze.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36538>
2025-08-05 22:55:14 +00:00
Marek Olšák
b769d5dcde nir: don't use variables as ralloc parents, use the shader instead
so that we can switch variables to gc_ctx

Reviewed-by: Timothy Arceri <tarceri@itsqueeze.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36538>
2025-08-05 22:55:13 +00:00
Marek Olšák
3c4a64e807 nir: eliminate most ralloc/malloc for nir_variable names
Store small names in a fixed-sized string in nir_variable.
GLSL IR does the same thing.

When compiling my shader-db with the gallium noop driver, it improves GLSL
compile times by 0.7% (much lower than anticipated).

For Unigine Heaven shaders:
- it eliminates 95.6% ralloc calls for nir_variable names
- the total number of ralloc calls is reduced by 11%

It also adds only 16B to nir_variable, while just the ralloc header
for the name would occupy 40B.

Reviewed-by: Timothy Arceri <tarceri@itsqueeze.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36538>
2025-08-05 22:55:12 +00:00
Marek Olšák
96ffc24e4e nir: add nir_variable_{set,append,steal}_name{f}() to modify nir_variable names
Setting variable names currently always uses ralloc, but the new
nir_variable_* helpers will mostly eliminate ralloc/malloc in a later
commit.

This just updates all places that touch nir_variable names to use the new
helpers.

Reviewed-by: Timothy Arceri <tarceri@itsqueeze.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36538>
2025-08-05 22:55:12 +00:00
Alyssa Rosenzweig
3e8575c037 nir,agx: pull lower_printf_buffer into backend
no other users now.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Marek Olšák <maraeo@gmail.com>
Reviewed-by: Mary Guillemard <mary.guillemard@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36516>
2025-08-03 21:27:50 +00:00
Alyssa Rosenzweig
1c28fc0a86 nir: add nir_inline_sysval pass
a bunch of drivers have versions of this, might as well make a common one.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Marek Olšák <maraeo@gmail.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Mary Guillemard <mary.guillemard@collabora.com>
Reviewed-by: John Anthony <john.anthony@arm.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36516>
2025-08-03 21:27:47 +00:00
Emma Anholt
d5826506ce nir,agx: Move AGX's loop (generalized) to shared NIR code.
When I went to use opt_reassociate for tu, I was advised that you want to
do this loop to get the best results.  If everyone needs it, let's make it
common code and explain what's going on.

In the process, also make it skip work appropriately when there's no
progress.

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36342>
2025-08-03 20:58:28 +00:00
Emma Anholt
d3ada77a6a nir: Move ST's force-persample-shading NIR pass to shared code.
This is about to grow a little.

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36496>
2025-08-03 20:27:39 +00:00