Commit graph

4613 commits

Author SHA1 Message Date
Christian Gmeiner
fb48d3d1da nir: add enta specific intrinsic used for txs lowering
Non of the know etnaviv GPUs support this feature in hardware
and the binary blob provides sizes via uniforms too.

Signed-off-by: Christian Gmeiner <cgmeiner@igalia.com>
Acked-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24217>
2023-07-21 08:52:03 +00:00
Christian Gmeiner
019e5cbd39 nir/print: print instr pass_flags
From time to time it can be helpful to "see" the pass_flags.

Signed-off-by: Christian Gmeiner <cgmeiner@igalia.com>
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24234>
2023-07-20 18:03:47 +00:00
Caio Oliveira
97c79cdf19 nir: Use instructions_pass() for nir_fixup_deref_modes()
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24220>
2023-07-19 20:15:12 +00:00
Alyssa Rosenzweig
d0f0afc6a4 nir: Initialize workgroup_size in builder_init_simple_shader
It can't be 0 in Vulkan.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24158>
2023-07-17 19:53:49 +00:00
Alyssa Rosenzweig
4f0f76346e nir: Add nir_lower_tess_coord_z pass
Lowers tess_coord to tess_coord_xy and math. Based on ir3's version.

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/24159>
2023-07-17 17:31:52 +00:00
Alyssa Rosenzweig
9109830bb0 nir: Promote tess_coord_r600 to tess_coord_xy
This intrinsic (vec2 tess_coord) is generally useful for non-r600 backends.
Promote it.

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/24159>
2023-07-17 17:31:52 +00:00
Rhys Perry
5e4029bfe5 nir/tests: add test for unsigned_upper_bound with loop header phis
Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23990>
2023-07-17 14:45:21 +00:00
Rhys Perry
1139d870f3 nir/unsigned_upper_bound: fix phi(bcsel)
This was looking at the wrong sources. src0 is the condition.

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Konstantin Seurer <konstantin.seurer@gmail.com>
Fixes: 72ac3f6026 ("nir: add nir_unsigned_upper_bound and nir_addition_might_overflow")
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23990>
2023-07-17 14:45:21 +00:00
Alyssa Rosenzweig
9bcdc45ee7 nir: Devendor load_sample_mask
AGX will use this too for its MSAA lowerings.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24148>
2023-07-15 19:48:30 +00:00
Alyssa Rosenzweig
56d61d9a64 nir: Add fence_{pbe,mem}_to_tex(_pixel)_agx intrinsics
Read-after-write hazards require special handling on AGX, since image loads are
implemented with texturing. Add intrinsics to handle these hazards.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24148>
2023-07-15 19:48:30 +00:00
Alyssa Rosenzweig
d54aa28b97 nir/legacy: Fix handling of fsat(fabs)
Consider code like:

    32x4  %2 = @load_interpolated_input (%1, %0 (0x0)) (base=0, component=0, dest_type=float32, io location=VARYING_SLOT_VAR0 slots=1 mediump)  // Color
    32x4  %3 = fabs %2
    32x4  %4 = fsat %3
    32x4  %5 = fsin %4

The existing logic would incorrectly tell the backend that both fabs and fsat
could be folded, and then half the shader disappears. Whoops. Fix by stopping
the folding in this case. I choose to do this check in the fsat rather than the
fabs because it's more straightforward (1 source vs N uses) but it's somewhat
arbitrary.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Pavel Ondračka <pavel.ondracka@gmail.com>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24116>
2023-07-13 22:43:36 +00:00
Alyssa Rosenzweig
34fcf6d479 nir/legacy: Fix fneg(load_reg) case
Consider the IR:

   %0 = load_reg
   %1 = fneg %0
   %2 = ffloor %1
   %3 = bcsel .., .., %1

Because the fneg has both foldable and non-foldable users, nir/legacy does not
fold the fneg into the load_reg. This ensures that the backend correctly emits a
dedicated fneg instruction (with the load_reg folded in) for the bcsel to use.
However, because the chasing helpers did not previously take other uses of a
modifier into account, the helpers would fuse in the fneg to the ffloor. Except
that doesn't work, because the load_reg instruction is supposed to be
eliminated. So we end up with broken chased IR:

   1 = fneg r0
   2 = ffloor -NULL
   3 = bcsel, ..., 1

The fix is easy: only fold modifiers into ALU instructions if the modifiers can
be folded away. If we can't eliminate the modifier instruction altogether, it's
not necessarily beneficial to fold it anyway from a register pressure
perspective. So this is probably ok. With that check in place we get correct IR

   1 = fneg r0
   2 = ffloor 1
   3 = bcsel, ..., 1

Fixes carchase/230.shader_test under softpipe.

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/24116>
2023-07-13 22:43:36 +00:00
Alyssa Rosenzweig
a2d56c4c73 nir/lower_blend: Use util enums
This avoids the silly compiler versions. Some bits are slightly more
complicated, because they have to account for inverted enum values (rather than
a separate invert bit), but this is a LOT friendlier to drivers using the pass
and it makes the pass itself more readable.

The conversion functions in panfrost/panvk will go away momentarily.

Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Italo Nicola <italonicola@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24076>
2023-07-13 21:03:32 +00:00
Christian Gmeiner
f831883af6 nir/lower_tex: optimize offset lowering for has_texture_scaling
Generates much better code and even helps to beat a blob driver.

Signed-off-by: Christian Gmeiner <cgmeiner@igalia.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24054>
2023-07-12 10:03:06 +00:00
Christian Gmeiner
9383009809 nir: rename has_txs to has_texture_scaling
Convert it to an opt-in for backends to prefer and use nir_load_texture_scale
instead of txs for nir lowerings.

Signed-off-by: Christian Gmeiner <cgmeiner@igalia.com>
Suggested-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24054>
2023-07-12 10:03:06 +00:00
Christian Gmeiner
9ddedf4554 nir: rename intrinsic to have a more generic nameing
Signed-off-by: Christian Gmeiner <cgmeiner@igalia.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24054>
2023-07-12 10:03:06 +00:00
Alyssa Rosenzweig
fded7e7b66 nir: Remove nir_register-based unit tests
Non-SSA functionality will become obsolete after nir_register is removed, so
there's no need to keep the tests around, and they will interfere with the
nir_register de-clawing.

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/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
e96a9a1b71 nir: Remove nir_lower_regs_to_ssa
It is now unused, as all internal producers of registers have been switched over
to intrinsics and no drivers call it.

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/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
9eab1e7521 nir/lower_shader_calls: Convert to register intrinsics
Yet another internal use of nir_register that gets lowered back to SSA after the
pass. Easy enough to replace with intrinsic-based registers instead.

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/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
61010e5255 nir: Add lower_vec_to_regs pass
This is a variant of nir_lower_vec_to_movs that produces register intrinsics
(store_reg with write masks) instead of masked moves.

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/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
aea8a70200 nir: Add intrinsics version of locals_to_regs
This isn't so bad. I still duplicated the pass because it makes a lot easier to
have them coexist, switch users over one by one, and then garbage collect the
old when we're done.

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/23089>
2023-07-12 01:34:27 +00:00
Faith Ekstrand
ae0408be1b nir/from_ssa: Support register intrinsics
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23089>
2023-07-12 01:34:27 +00:00
Faith Ekstrand
29b2ace184 nir/from_ssa: Make additional assumptions in coalescing
At this point, everything is SSA.  Also, NIR no longer allows different
numbers of components on the two sides of a phi so we can just assert
rather than trying to gracefully handle mismatches.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
36b29201fa nir: Produce intrinsics in lower_{phis,ssa_defs}_to_regs
A number of passes lower SSA partially to registers, do work that would be
invalid in SSA, and then go back into SSA with nir_lower_regs_to_ssa. As a step
towards replacing nir_register with intrinsics,
the nir_lower_{phis,ssa_defs}_to_regs passes are changed to produce intrinsics
instead of nir_registers, and their callers are updated to call
nir_lower_reg_intrinsics_to_ssa instead of nir_lower_regs_to_ssa to compensate.

Jointly authored with Faith.

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/23089>
2023-07-12 01:34:27 +00:00
Faith Ekstrand
73e191924c nir: Add a reg_intrinsics flag to nir_convert_from_ssa
It doesn't do anything yet. We leave that to the subsequent patches so we can
keep the tree-wide refactor as simple as possible.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
bcf3a622d1 nir: Add new version of lower_regs_to_ssa
in the sense of operating on register intrinsics instead of nir_registers.

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/23089>
2023-07-12 01:34:27 +00:00
Alyssa Rosenzweig
2eb554af48 nir: Add legacy data structures & helpers
These are registerful versions of core nir_src/nir_dest which will become
SSA-only soon enough, and modifierful versions of nir_alu_src/nir_alu_dest.
The latter will let us remove modifiers from nir_alu_instr finally.

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/23089>
2023-07-12 01:34:26 +00:00
Alyssa Rosenzweig
d313eba94e nir: Add pass for trivializing register access
After running the pass, all register access intrinsics are guaranteed to be
"trivial" in the sense that the program is free of hazards preventing
propagating them away without inserting any copies.

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/23089>
2023-07-12 01:34:26 +00:00
Alyssa Rosenzweig
1d6c06e4b9 nir: Add helpers for walking register uses
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/23089>
2023-07-12 01:34:26 +00:00
Alyssa Rosenzweig
7229bffcb1 nir: Add intrinsics for register access
Note the writemask handling is chosen for consistency with the rest of NIR. In
every other instance, writemask=w requires a vec4 source. This is hardcoded into
nir_validate and nir_print as what it means to have a writemask.

More importantly, consistency with how register writemasks currently work.
nir_print hides it, but r0.w = fneg ssa_1.x is actually a vec4 instruction with
source ssa_1.xxxx. As a silly example nir_dest_num_components(that) = 4 in the
old model. I realize this is quite strange coming from a scalar ISA, but it's
perfectly natural for the class of vec4 hardware for which this was designed. In
that hardware, conceptually all instructions are vec4`, so the sequence "fneg
ssa_1 and write to channel w" is implemented as "fneg a vec4 with ssa_1.x in the
last component and write that vec4 out but mask to write only the w channel".

Isn't this inefficient? It can be. To save power, Midgard has scalar ALUs in
addition to vec4 ALUs. Those details are confined to the backend VLIW scheduler;
the instruction selection is still done as vec4. This mechanism has little in
common with AMD's SALUs. Midgard has a wave size of 1, with special hacks for
derivatives.

As a result, all backends consuming register writemasks are expecting this
pattern of code. Changing the store to take a vec1 instead of a vec4 would
require changing every backend to reswizzle the sources to resurrect the vec4. I
started typing a branch to do this yesterday, but it made a mess of both Midgard
and nir-to-tgsi. Without any good reason to think it'd actually help
performance, I abandoned the idea. Getting all 15 backends converted to the
helpers is enough of a challenge without forcing 10 backends to reswizzle their
sources too.

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/23089>
2023-07-12 01:34:26 +00:00
Konstantin Seurer
f160f7c525 nir/opt_dead_cf: Clarify comment
Make it obvious that the comment is about the block stitching behavior
of nir_cf_node_remove.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22064>
2023-07-11 17:32:55 +00:00
Konstantin Seurer
4b5737b785 nir/opt_dead_cf: Run dead_cf_block while it makes progress
Previously, nir_opt_dead_cf could skip dead CF nodes because overwriting
cur after dead_cf_block is not enough to cover the whole CF list.
foreach_list_typed would select the next node, skipping the node that
previously made progress:

block 1
if (true) {}
block 2
if (true) {}
block 3
if (true) {}

Would turn into:

block 1, then, block 2
if (true) { }
block 3, then

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22064>
2023-07-11 17:32:55 +00:00
Konstantin Seurer
6532751e4d nir/lower_shader_calls: Remat derefs after shader calls
This avoids spilling deref instructions by wrapping shader calls inside
dummy blocks, rematerializing derefs in their use blocks and removing
the dummy blocks.

Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22064>
2023-07-11 17:32:55 +00:00
Yonggang Luo
97014036a9 nir: Remove nir_builder_init, it's not used anymore
Signed-off-by: Yonggang Luo <luoyonggang@gmail.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24038>
2023-07-10 19:20:18 +00:00
Yonggang Luo
48a25ef700 treewide: Remove all usage of nir_builder_init with nir_builder_create and nir_builder_at
Signed-off-by: Yonggang Luo <luoyonggang@gmail.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24038>
2023-07-10 19:20:17 +00:00
Italo Nicola
65d6f5aed2 nir: add options to lower y_vu, yv_yu, yx_xvxu and xy_vxux
`y_vu` will be used to convert NV21 to RGB.
`yv_yu` will be used to convert YVYU and VYUY to RGB when the
subsampling formats PIPE_FORMAT_R8B8_R8G8 and PIPE_FORMAT_B8R8_G8R8
are supported.
`yx_xvxu` and `xy_vxux` will be used to convert YVYU and VYUY to RGB
when those subsampling formats are not supported.

Signed-off-by: Italo Nicola <italonicola@collabora.com>
Reviewed-by: Daniel Stone <daniels@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21219>
2023-07-10 16:29:13 +00:00
Erik Faye-Lund
1f9a3ee011 nir: use nir_intrinsic_get_var
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24004>
2023-07-10 16:06:40 +02:00
Erik Faye-Lund
67a7b66b90 nir: constify intrin
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24004>
2023-07-10 16:06:40 +02:00
Lionel Landwerlin
0a6dd797ea nir/opt_shrink_vectors: enable sparse intrinsics shrinking
Image sparse loads can be stripped from their sparse component if
unused and turned into non sparse variants.

Texture sparse accesses can also be turned off if unused.

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23995>
2023-07-06 13:16:13 +00:00
Timothy Arceri
7cff8d85c0 nir/glsl: add nir_var_declared_implicitly enum
This will be used in the following patch.

Reviewed-by: Emma Anholt <emma@anholt.net>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22846>
2023-07-06 01:21:06 +00:00
Alyssa Rosenzweig
0e7e6f2a0d nir: Fix breaking in nir_foreach_phi(_safe)
When I reading through some of my older commits I noticed that `break` in
`nir_foreach_phi` is broken because I used the two-loop trick wrong. Rewrite the
macros to fix this, and also to generally be a lot cleaner.

Fixes: 7dc297cc14 ("nir: Add nir_foreach_phi(_safe) macro")
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23957>
2023-07-05 08:42:23 -04:00
Marcin Ślusarz
3d7513ee8e nir: add cheap shortcut for wg id to wg idx lowering
... for platforms where integer division is expensive

Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22334>
2023-07-04 09:15:08 +00:00
Marcin Ślusarz
e7ca9d70f5 nir: lower num_workgroups to constants
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22334>
2023-07-04 09:15:08 +00:00
Marcin Ślusarz
b9eeee8554 nir: use constant components of num_workgroups in wg id to wg idx lowering
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22334>
2023-07-04 09:15:08 +00:00
Marcin Ślusarz
7ebfbc97a8 nir: use wg id to wg idx shortcut if two dims of num_workgroups are 1
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22334>
2023-07-04 09:15:07 +00:00
Marcin Ślusarz
b5792c1a34 nir: extract try_lower_id_to_index_1d
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/22334>
2023-07-04 09:15:07 +00:00
Caio Oliveira
608504c774 nir/print: Reformat the preds/succs block information
- Always print preds in same line as block name;
- Use a single line for empty blocks;
- Align preds/succs with the instructions.

```
if %29 {
    block b4:    // preds: b3
    32     %30 = load_const (0x00000000 = 0.000000)
    32x4   %31 = @vulkan_resource_index (%30 (0x0)) (desc_set=0, binding=0, desc_type=SSBO)
    32x4   %32 = @load_vulkan_descriptor (%31) (desc_type=SSBO)
    32x4   %33 = deref_cast (Storage *)%32 (ssbo Storage)  (ptr_stride=0, align_mul=4, align_offset=0)
    32x4   %34 = deref_struct &%33->fail (ssbo uint)  // &((Storage *)%32)->fail
    32     %36 = @deref_atomic (%34, %35 (0x1)) (access=1, atomic_op=iadd)
                 // succs: b6
} else {
    block b5:  // preds: b3, succs: b6
}
```

Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Acked-by: Kenneth Graunke <kenneth@whitecape.org>
Acked-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23564>
2023-07-03 22:18:07 +00:00
Caio Oliveira
a188337972 nir/print: Print div/con annotation first
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Acked-by: Kenneth Graunke <kenneth@whitecape.org>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23564>
2023-07-03 22:18:07 +00:00
Caio Oliveira
884debdee3 nir/print: Use 4-space indentation
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Acked-by: Kenneth Graunke <kenneth@whitecape.org>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23564>
2023-07-03 22:18:06 +00:00
Caio Oliveira
9215aad7da nir/print: Use // for comments
Makes it easier to copy snippets of shaders into code or
test comments without worrying about conflict with `/* */`.

Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Acked-by: Kenneth Graunke <kenneth@whitecape.org>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/23564>
2023-07-03 22:18:06 +00:00