Commit graph

298 commits

Author SHA1 Message Date
Marek Olšák
b2d32ae246 nir: add nir_intrinsic_load_per_primitive_input, split from io_semantics flag
Instead of having 1 bit in nir_io_semantics indicating a per-primitive
FS input, add a dedicated intrinsic for it.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/29895>
2024-07-23 16:13:16 +00:00
Daniel Schürmann
7af16e9f1e nir/shader_info: remove uses_demote
This flag is mostly redundant with uses_discard and was only
introduced to implement demote with LLVM when it didn't have
that intrinsic.

Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27617>
2024-06-17 19:37:16 +00:00
Faith Ekstrand
290cbf413c nir/print: Improve divergence information
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/29591>
2024-06-13 20:43:44 +00:00
Alyssa Rosenzweig
2912f531a7 nir: add texops for AGX border colour emulation
AGX has limited border colour hardware. To support full
customBorderColorWithoutFormat semantics, we're forced to emulate in shaders at
a substantial performance penalty. Actually, that's needed just to pass CTS
because of other hardware issues stacking on top of each others... Hooray!

Add the texops we need to facilitate efficient custom border colour lowering.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Acked-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/29179>
2024-05-14 04:57:24 +00:00
Saroj Kumar
221371e903 mesa: replace shader_info::source_sha1
Replace shader_info::source_sha1 with shader_info::source_blake3 in compiler, mesa and radeonsi.

Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Acked-by: David Heidelberg <david.heidelberg@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28156>
2024-05-09 20:08:18 +00:00
Marek Olšák
1a791c1303 nir: add nir_atomic_op_ordered_add_gfx12_amd
for streamout

Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Acked-By: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28889>
2024-04-30 17:17:25 +00:00
Marek Olšák
d4cfcbdde8 nir: add ACCESS_CP_GE_COHERENT_AMD
required by amd gfx12

Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Acked-By: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28889>
2024-04-30 17:17:25 +00:00
Konstantin Seurer
ea863c0c1c nir/print: Do not access invalid indices of load_uniform
load_uniform does not have io_semantics and component.

Fixes: a83fd26 ("nir/print: stop trying to match i/o vars using base/driver_location")
Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Reviewed-by: Christian Gmeiner <cgmeiner@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28962>
2024-04-28 16:06:46 +02:00
Mike Blumenkrantz
cb597cb85e nir/print: print io instr->name if available
this will always be more accurate than trying to find the name from
a variable

Acked-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/28814>
2024-04-24 12:35:59 +00:00
Mike Blumenkrantz
a83fd26d0e nir/print: stop trying to match i/o vars using base/driver_location
this is broken

Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28552>
2024-04-18 15:30:47 +00:00
Mike Blumenkrantz
6ae2147dd6 nir: print i/o variables in location order
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25036>
2024-04-17 00:46:40 +00:00
Timur Kristóf
723b3d354e nir/print: Print per-primitive and explicit strict IO info.
Signed-off-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28685>
2024-04-14 19:51:11 +00:00
Konstantin Seurer
1ff8659f41 nir/print: Fix printing booleans with bit_size>1
Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28187>
2024-04-09 07:13:01 +00:00
Faith Ekstrand
b069151e62 nir/print: Inline print_ssa_use()
It has one caller.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28300>
2024-03-25 15:55:48 +00:00
Alyssa Rosenzweig
49a89911c4 nir/print: do not print empty lists on intrinsics
before:

  32     %0 = @load_vertex_id () ()

after:

   32     %0 = @load_vertex_id

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27925>
2024-03-12 19:00:26 +00:00
Marek Olšák
813f37a8ed nir: add nir_block::divergent to indicate a divergent entry condition
to be used by nir_opt_varyings

Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28049>
2024-03-12 00:29:03 +00:00
Marek Olšák
936690f733 nir: print nir_io_semantics::invariant
this was missing

Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/28049>
2024-03-12 00:29:03 +00:00
Lionel Landwerlin
259cdc5496 nir: add additional flag to resource_intel for embedded samplers
This will enable specific lowering of embedded samplers.

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Ivan Briano <ivan.briano@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22151>
2024-02-29 07:05:06 +00:00
Sagar Ghuge
c984d6e2fc nir: Drop intel specific lowering code
In previous patches, we have moved the Intel specific lowering code in
brw_nir_lower_texture file. We can go ahead and drop the Intel specific
texture source too.

Signed-off-by: Sagar Ghuge <sagar.ghuge@intel.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27458>
2024-02-12 21:25:48 +00:00
Ian Romanick
c8ba2bc2f0 nir: Pack texture LOD and array index to a single 32-bit value
v2: Fix clamped_ai calculation in nir_lower_tex.c. Add
nir_tex_src_combined_lod_and_array_index_intel to
print_tex_instr. Suggested by Sagar.

Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27305>
2024-02-02 02:39:10 +00:00
Konstantin Seurer
e3c2dc2324 nir/print: Rename workgroup-size to workgroup_size
Every other field uses _ instead of -.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27318>
2024-01-30 21:19:40 +00:00
Konstantin Seurer
449e44d6d3 nir/print: Don't print shared_size twice
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27318>
2024-01-30 21:19:40 +00:00
Marek Olšák
b6e98677c3 nir/print: print PATCH0 and VARn_16BIT names instead of numbers for TCS and TES
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26275>
2023-11-25 09:50:19 -05:00
Daniel Schürmann
f1110576d9 nir: add info.fs.require_full_quads
This flag indicates the requirement of helper invocations
in fragment shaders, independent from any present instructions.
This fixes the lowering of OpGroupNonUniformQuad* instructions.

Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26026>
2023-11-22 11:31:52 +01:00
Daniel Schürmann
1179d83a89 nir: remove info.fs.needs_all_helper_invocations
Use info.uses_wide_subgroup_intrinsics instead.

Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26026>
2023-11-22 11:31:11 +01:00
Alyssa Rosenzweig
8d9d9d0207 nir/print: handle adjacency
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Antonino Maniscalco <antonino.maniscalco@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26056>
2023-11-07 00:05:54 +00:00
Alyssa Rosenzweig
6014f745d5 nir,vtn: Add exported bool to nir_function
For optimizing libraries.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25498>
2023-11-02 11:37:46 +00:00
Faith Ekstrand
1a2e8290ab nir: Add NV-specific texture opcodes
These are for implementing various texture queries.

Acked-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25000>
2023-10-24 22:21:18 +00:00
Marek Olšák
f3886e9c02 nir: split FLOAT_CONTROLS_SIGNED_ZERO_INF_NAN_PRESERVE_FP* flags
GLSL doesn't preserve NaNs, but it optionally preserves Infs.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25392>
2023-10-17 17:27:12 +00:00
Emma Anholt
40f22f8ecd nir/print: Decode system values in the variable declarations.
decl_var system INTERP_MODE_NONE none vec4 #0
decl_var system INTERP_MODE_FLAT none mediump uint #1

turns into:

decl_var system INTERP_MODE_NONE none vec4 #0 (SYSTEM_VALUE_FRAG_COORD)
decl_var system INTERP_MODE_FLAT none mediump uint #1 (SYSTEM_VALUE_SUBGROUP_INVOCATION)

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25647>
2023-10-12 22:52:42 +00:00
Alyssa Rosenzweig
569d44eff4 nir/print: Handle KERNEL
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25625>
2023-10-12 21:03:31 +00:00
Alyssa Rosenzweig
3325b4778a nir: Add ACCESS_CAN_SPECULATE
Determining whether it is safe to hoist a load instruction out of control flow
depends on complex hardware and driver details. Rather than encoding this as
knobs in every NIR pass that wants to do so (notably nir_opt_preamble and
nir_opt_peephole_select), add a per-load ACCESS flag for backends to set.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Connor Abbott <cwabbott0@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24011>
2023-10-10 13:51:00 +00:00
Marek Olšák
b1bbe4e190 nir: gather dual slot input information
Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25394>
2023-10-07 11:18:40 +00:00
Marek Olšák
0f2491cbdd nir: add dual-slot input information into load_input intrinsics
This is necessary to allow optimizing VS inputs after nir_lower_io, which
is currently impossible because the loss of dual-slot information in NIR
would break VS inputs. With this, driver locations can be recomputed by
calling nir_recompute_io_bases. It's a prerequisite for optimizing varyings
with lowered IO.

When this is used, we will be able to eliminate unused dual-slot VS inputs
as well as unused low and high halves of dual-slot VS inputs for the first
time, which can happen due to optimizations of varyings. Without this,
st/mesa binds vertex buffers for dual-slot inputs that are fully or
partially unused in the shader.

Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25394>
2023-10-07 11:18:40 +00:00
Caio Oliveira
3105d516d0 nir: Add new intrinsics for Cooperative Matrix
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Reviewed-by: Bas Nieuwenhuizen <bas@basnieuwenhuizen.nl>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23825>
2023-09-28 07:35:02 +00:00
Dave Airlie
8a95b43880 spirv/nir: parse function control and store in nir.
This just lets the nir access the inline/dont inline attributes

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24687>
2023-09-12 01:57:50 +00:00
Alyssa Rosenzweig
8c7629524e nir/print: Print access qualifiers for intrinsics
Instead of printing an opaque integer that needs to be manually decoded.
Example output:

    32x4   %7 = @image_load (%4 (0x0), %6, %5 (0x0), %4 (0x0)) (image_dim=2D, image_array=false, format=r8g8b8a8_snorm, access=readonly|reorderable, range_base=0, dest_type=float32)

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24798>
2023-08-22 20:24:40 +00:00
Konstantin Seurer
ccc52ae887 nir: Add shader enqueue data structures and handling
There are two new variable modes:
- nir_var_mem_node_payload
- nir_var_mem_node_payload_in

Also add a few more intrinsics and some shader info.

Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24512>
2023-08-18 16:57:22 +00:00
Faith Ekstrand
65b6ac8aa4 nir: Rename nir_instr_type_ssa_undef to nir_instr_type_undef
We already renamed the type, we just need to rename the enum and the
casting helper functions.

Generated with sed:

    sed -i -e 's/nir_instr_type_ssa_undef/nir_instr_type_undef/g' src/**/*.h src/**/*.c src/**/*.cpp
    sed -i -e 's/nir_instr_as_ssa_undef/nir_instr_as_undef/g' src/**/*.h src/**/*.c src/**/*.cpp

and two tiny whitespace fixups in lima.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24703>
2023-08-15 17:44:27 +00:00
Faith Ekstrand
4695bebc79 nir: Drop nir_dest
Instead, we replace every use of it with nir_def.  Most of this commit
was generated by sed:

   sed -i -e 's/dest.ssa/def/g' src/**/*.h src/**/*.c src/**/*.cpp

A few manual fixups were required in lima and the nir_legacy code.

Acked-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24674>
2023-08-14 21:22:53 +00:00
Faith Ekstrand
6c1d32581a nir: Drop nir_alu_dest
Instead, we replace it directly with nir_def.  We could replace it with
nir_dest but the next commit gets rid of that so this avoids unnecessary
churn.  Most of this commit was generated by sed:

   sed -i -e 's/dest.dest.ssa/def/g' src/**/*.h src/**/*.c src/**/*.cpp

There were a few manual fixups required in the nir_legacy.c and
nir_from_ssa.c as nir_legacy_reg and nir_parallel_copy_entry both have a
similar pattern.

Acked-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24674>
2023-08-14 21:22:53 +00:00
Faith Ekstrand
0dd76b1abb nir/print: Replace all dest printing with print_def
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24674>
2023-08-14 21:22:53 +00:00
Alyssa Rosenzweig
09d31922de nir: Drop "SSA" from NIR language
Everything is SSA now.

   sed -e 's/nir_ssa_def/nir_def/g' \
       -e 's/nir_ssa_undef/nir_undef/g' \
       -e 's/nir_ssa_scalar/nir_scalar/g' \
       -e 's/nir_src_rewrite_ssa/nir_src_rewrite/g' \
       -e 's/nir_gather_ssa_types/nir_gather_types/g' \
       -i $(git grep -l nir | grep -v relnotes)

   git mv src/compiler/nir/nir_gather_ssa_types.c \
          src/compiler/nir/nir_gather_types.c

   ninja -C build/ clang-format
   cd src/compiler/nir && find *.c *.h -type f -exec clang-format -i \{} \;

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Acked-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Acked-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24585>
2023-08-12 16:44:41 -04:00
Faith Ekstrand
777d336b1f nir: clang-format src/compiler/nir/*.[ch]
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24382>
2023-08-12 19:27:28 +00:00
Alyssa Rosenzweig
10cdc0ad9f nir: Add load_coefficients_agx intrinsic
For lowering interpolation.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24498>
2023-08-11 09:50:11 +00:00
Mike Blumenkrantz
f75ba983ca nir/print: always group variables by type when printing
Acked-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23752>
2023-08-08 13:51:34 +00:00
Mike Blumenkrantz
4a783abd79 nir/print: print location names for (some) tess slots
these should be fine to print

Acked-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23752>
2023-08-08 13:51:34 +00:00
Alyssa Rosenzweig
95e3df39c0 treewide: sed out more is_ssa
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24432>
2023-08-03 22:40:28 +00:00
Alyssa Rosenzweig
efc1c3261a nir/print: Drop legacy NIR
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24432>
2023-08-03 22:40:28 +00:00
Alyssa Rosenzweig
5fead24365 treewide: Drop is_ssa asserts
We only see SSA now.

Via Coccinelle patch:

    @@
    expression x;
    @@

    -assert(x.is_ssa);

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Acked-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24432>
2023-08-03 22:40:28 +00:00