Commit graph

9179 commits

Author SHA1 Message Date
Timothy Arceri
99a80ac930 glsl: add half float geometric functions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:14 +00:00
Timothy Arceri
6a170051a9 glsl: add support for half float packing functions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:14 +00:00
Timothy Arceri
c386d56915 glsl: add half float support for common functions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:14 +00:00
Timothy Arceri
eea1c1fa7b glsl: add f2f16() helper to ir_builder
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:14 +00:00
Timothy Arceri
94865a45f8 glsl: add half float exponential functions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:14 +00:00
Timothy Arceri
829c99def1 glsl: add half float angle and trigonometry functions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
3ff22ee4ae glsl: add some new helpers for half float builtin functions
These helpers will be used in the following patches.

Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
66633dbbd0 glsl: add ubo packing support for half floats
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
5803417c67 glsl: add implicit half float conversions
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
b83477973f glsl: add glsl_type_is_float_16() helper
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
5eea4c42f5 glsl: update assert to allow for half float support
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
b2a7717717 glsl: add explicit half float conversion support
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
edbe1b5bb4 glsl: add half float support to the parser
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Timothy Arceri
72fb49516e glsl: add basic params for AMD_gpu_shader_half_float extension
This sets us the basic infrastructure to required to add an
extension.

Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18540>
2024-02-12 13:23:13 +00:00
Alyssa Rosenzweig
9c006d5dce nir/passthrough_gs: flesh out gs_in_prim
geometry shaders don't specify the input topology, only the class of topology.
normalize when generating a passthrough gs.

asahi will be more picky about this in the future.

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/27457>
2024-02-09 11:53:31 +00:00
Alyssa Rosenzweig
3c3f7cb747 compiler,gallium: move u_decomposed_prim to common
nir will use

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/27457>
2024-02-09 11:53:31 +00:00
Daniel Schürmann
312d0784c5 spirv: implement SPV_KHR_quad_control
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27277>
2024-02-09 05:32:35 +00:00
Daniel Schürmann
6588f5a123 spirv: implement SPV_KHR_maximal_reconvergence
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27277>
2024-02-09 05:32:35 +00:00
Christian Gmeiner
7c198d4404 isaspec: encode.py: Include assert.h
pack_field(..) makes use of assert(..).

Signed-off-by: Christian Gmeiner <christian.gmeiner@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27491>
2024-02-06 16:38:01 +00:00
Marek Olšák
e98bbcad17 nir: add vertex divergence into nir_divergence_analysis
This is a prerequisite for the new nir_opt_varyings pass.
It reuses the same divergent field in nir_def and nir_loop.

Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:52 -05:00
Marek Olšák
5ffa4d879c nir: add a lower_mediump_io callback into options
This will be called by the GLSL linker before nir_opt_varyings.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:51 -05:00
Marek Olšák
ecf0fe09f0 nir: replace lower_io_variables with a GLSL NIR flag
This stops using it in nir_lower_io_passes because all callers call it
only when it's true.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:49 -05:00
Marek Olšák
c4acab77a8 nir: remove and replace underused option pack_varying_options
This will also be used by nir_opt_varyings.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:47 -05:00
Marek Olšák
c844b5dc85 nir: relax validation failure for generic TCS outputs with no_varying
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:45 -05:00
Marek Olšák
d84a616d4d nir: remove INTERP_MODE_COLOR
It's only used by radeonsi and doesn't have to be public.

Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26918>
2024-02-02 16:45:35 -05:00
Job Noorman
60413e11c2 ir3: optimize subgroup operations using brcst.active
Follow the blob and optimize subgroup operation using brcst.active and
getlast when supported.

The transformation consists of two parts. First, a NIR transform
replaces subgroup operations with a sequence of new brcst_active_ir3
intrinsics followed by a new [type]_clusters_ir3 intrinsic (where type
can be reduce, inclusive_scan, or exclusive_scan).

The brcst_active_ir3 intrinsic is lowered directly to a brcst.active
instruction. The other intrinsics get lowered to a new macro
(OPC_SCAN_CLUSTERS_MACRO) which later gets emitted as a loop (using
getlast/getone) that iterates all clusters and produces the requested
scan result.

OPC_SCAN_CLUSTERS_MACRO has a number of optional arguments. First, since
the exclusive scan result is not a natural by-product of the loop but
has to be calculated explicitly, its destination is optional. This is
necessary since adding it unconditionally will produce unused
instructions that won't be DCE'd anymore at this point. Second, when
performing 32b MUL_U reductions (that expand to multiple instructions),
an extra scratch register is necessary.

Closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/6387
Signed-off-by: Job Noorman <jnoorman@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/26950>
2024-02-02 19:49:22 +00:00
Konstantin Seurer
c925b6019d radv/rt: Lower ray payloads like hit attribs
Reviewed-by: Friedrich Vock <friedrich.vock@gmx.de>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27051>
2024-02-02 16:36:15 +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
Timothy Arceri
bc0178af57 glsl: don't tree graft globals
As per this optimisations description:

"Takes assignments to variables that are dereferenced only
once and pastes the RHS expression into where the variables
dereferenced."

However the optimisation is run at compile time before multiple
shaders from the same stage could have been pasted together.
So this optimisation can incorrectly assume a global is only
referenced once since it cannot see the other pieces of the
shader stage until link time.

Here we skip the optimisation if the variable is a global. We
could change it to only run at link time however this
optimisation is only run at link time if we are being forced
to use GLSL IR to inline a function that glsl to nir cannot
handle and this will also be removed in a future patchset.

Closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/10482
Fixes: d75a36a9ee ("glsl: remove do_copy_propagation_elements() optimisation pass")

Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27351>
2024-02-01 23:15:24 +00:00
Timothy Arceri
ec4990ef41 glsl: remove now unused GLSL IR validate_geometry_shader_emissions()
We now use the nir version of this validation pass instead.

Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27377>
2024-02-01 01:53:15 +00:00
Timothy Arceri
ce3420394c glsl: use nir version of geom stream validation
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27377>
2024-02-01 01:53:15 +00:00
Timothy Arceri
f21e2e6229 glsl: add nir version of validate_geometry_shader_emissions()
We will use this to replace the GLSL IR version.

Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27377>
2024-02-01 01:53:15 +00:00
Friedrich Vock
321e2cee53 vtn: Use secure_getenv for shader dumping
Reviewed-by: Eric Engestrom <eric@igalia.com>
Cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27381>
2024-01-31 17:28:12 +00:00
Antoine Coutant
445aacb421 clc: retrieve libclang path at runtime.
LLVM_LIB_DIR is a variable used for runtime compilations.
When cross compiling, LLVM_LIB_DIR must be set to the
libclang path on the target. So, this path should not
be retrieved during compilation but at runtime.

dladdr uses an address to search for a loaded library.
If a library is found, it returns information about it.
The path to the libclang library can therefore be
retrieved using one of its functions. This is useful
because we don't know the name of the libclang library
(libclang.so.X or libclang-cpp.so.X)

v2 (Karol): use clang::CompilerInvocation::CreateFromArgs for dladdr
v3 (Karol): follow symlinks to fix errors on debian

Fixes: e22491c832 ("clc: fetch clang resource dir at runtime")
Signed-off-by: Antoine Coutant <antoine.coutant@smile.fr>
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by (v1): Jesse Natalie <jenatali@microsoft.com>

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25568>
2024-01-31 12:23:49 +00:00
Karol Herbst
8efd11fce9 clc: force fPIC for every user when using shared LLVM
As we want to start using `dladdr`, this is needed to prevent `dladdr`
returning information of the wrong file.

Fixes tag as it's required by the actual fix.

Signed-off-by: Karol Herbst <kherbst@redhat.com>
Fixes: e22491c832 ("clc: fetch clang resource dir at runtime")
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25568>
2024-01-31 12:23:49 +00:00
Karol Herbst
37a1346347 meson: remove opencl-external-clang-headers option and rely on shared-llvm
Setting opencl-external-clang-headers to enabled while using shared LLVM
was broken and this option was mostly used for windows to force static
inclusion of opencl base headers.

Simply relying on the shared-llvm option here is enough to get what we
want.

Signed-off-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/25568>
2024-01-31 12:23:49 +00:00
Daniel Schürmann
faf3bb644d spirv: Update headers and grammar JSON
1.3.276.0

Acked-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27273>
2024-01-30 22:51:52 +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
Gert Wollny
0ab3b3c641 nir/builder: Fix compilation with gcc-13 when tsan is enabled
../src/compiler/nir/nir_builder.h: In function ‘nir_build_deref_follower’:
../src/compiler/nir/nir_builder.h:1607:1: error: control reaches end of non-void function [-Werror=return-type]
 1607 | }

Fixes: 4a4e175738
    nir: Support deref instructions in lower_var_copies

Signed-off-by: Gert Wollny <gert.wollny@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27345>
2024-01-30 20:42:07 +00:00
Gert Wollny
80a1b91601 nir/lower_int64: Fix compilation with gcc-13 and tsan enabled
../src/compiler/nir/nir_lower_int64.c: In function ‘lower_int64_intrinsic’:
../src/compiler/nir/nir_lower_int64.c:1347:1: error: control reaches end of non-void function [-Werror=return-type]
1347 | }

Fixes: bf7a114246
   nir/lower_int64: Add lowering for some 64-bit subgroup ops

Signed-off-by: Gert Wollny <gert.wollny@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27345>
2024-01-30 20:42:07 +00:00
Daniel Schürmann
5df7be8017 spirv: Fix SpvOpExpectKHR
This instruction behaves the same as *OpCopyObject* by making a copy of _Value_.

Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27265>
2024-01-30 19:09:42 +00:00
Faith Ekstrand
48ebfeba34 nak: Add a source barrier intrinsic
This just inserts a GPU stall until the given source is available.  We
need this in order to properly implement shader clock.

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27303>
2024-01-26 16:55:50 +00:00
Jesse Natalie
86b5c9278c glsl: Work around MSVC arm64 optimizer bug
Reviewed-by: Yonggang Luo <luoyonggang@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27199>
2024-01-26 00:22:53 +00:00
Friedrich Vock
9f22b95956 nir: Handle casts in nir_opt_copy_prop_vars
Cc: mesa-stable

Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27197>
2024-01-24 12:39:48 +00:00
Friedrich Vock
6c845ed548 nir: Make is_trivial_deref_cast public
Cc: mesa-stable

Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27197>
2024-01-24 12:39:48 +00:00
Rhys Perry
e465ac2561 nir/lower_shader_calls: remove CF before nir_opt_if
Otherwise, opt_if_simplification() can attempt to insert an inot after a
jump.

Fixes RADV compilation of a Cyberpunk 2077 pipeline with
PIPELINE_CREATE_DISABLE_OPTIMIZATION_BIT.

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27193>
2024-01-23 19:02:03 +00:00
Rhys Perry
015b0d678f nir/lower_non_uniform: set non_uniform=false when lowering is not needed
Fixes RADV compilation of a Doom Eternal pipeline with
PIPELINE_CREATE_DISABLE_OPTIMIZATION_BIT, because
nir_opt_non_uniform_access was skipped and later passes don't expect
non-uniform access.

Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Fixes: b1619109ca ("nir/lower_non_uniform: remove non_uniform flags after lowering")
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27192>
2024-01-23 18:09:39 +00:00
Karol Herbst
f2b7c4ce29 nir: rework and fix rotate lowering
No driver supports urol/uror on all bit sizes. Intel gen11+ only for 16
and 32 bit, Nvidia GV100+ only for 32 bit. Etnaviv can support it on 8,
16 and 32 bit.

Also turn the `lower` into a `has` option as only two drivers actually
support `uror` and `urol` at this momemt.

Fixes crashes with CL integer_rotate on iris and nouveau since we emit
urol for `rotate`.

v2: always lower 64 bit

Fixes: fe0965afa6 ("spirv: Don't use libclc for rotate")
Signed-off-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by (Intel and nir): Ian Romanick <ian.d.romanick@intel.com>

Reviewed-by: David Heidelberg <david.heidelberg@collabora.com>
Acked-by: Yonggang Luo <luoyonggang@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27090>
2024-01-22 10:27:44 +00:00
Georg Lehmann
d641750573 nir: add lowering for boolean shuffle
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/27116>
2024-01-19 20:13:34 +00:00