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
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
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
d029686e20
aco/isel: fix output args init stack buffer overflow
...
BITSET range functions include the end of the range.
Fixes: eb249bb18ependingchaos02@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37361 >
2025-09-16 09:24:03 +00:00
849344dc08
aco: set program->dev.fused_mad_mix=true for GFX940
...
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35655 >
2025-09-16 07:02:32 +00:00
8c4f0b1353
aco: gfx940 has no mad f32 instruction
...
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35655 >
2025-09-16 07:02:32 +00:00
8803388d15
ci: Update to Debian 13 (trixie)
...
Switch containers from Debian 12 (bookworm) to Debian 13 (trixie).
Trixie ships LLVM 19 by default, so we no longer need to add LLVM repos
to install llvm-19.
Notably, trixie also uses Python 3.13.
Closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/6994
Signed-off-by: Valentine Burley <valentine.burley@collabora.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35853 >
2025-09-16 06:16:21 +00:00
b16c62b6b2
meson: Relax -Wmaybe-uninitialized errors
...
Suggested-by: @eric
Signed-off-by: Valentine Burley <valentine.burley@collabora.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/35853 >
2025-09-16 06:16:20 +00:00
0bc0ead674
radv: set DRLR mapping info from inheritance info when present
...
These two structs are allowed to be in pNext and they should match
the primary command buffer info.
Found while implementing a new extension.
Cc: mesa-stable
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37278 >
2025-09-15 19:29:34 +00:00
5907dbfc09
radv: remove redundant RADV_DYNAMIC_RASTERIZATION_SAMPLES
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:42 +00:00
2084cb59f2
radv: remove redundant RADV_DYNAMIC_POLYGON_MODE
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:42 +00:00
c1a1aed665
radv: remove redundant RADV_DYNAMIC_LINE_RASTERIZATION_MODE
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:41 +00:00
d8bc573ee9
radv: remove redundant RADV_DYNAMIC_PRIMITIVE_TOPOLOGY
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:40 +00:00
43d7795274
radv: pre-compute vgt_outprim_type
...
This will allow us to optimize the number of states to emit.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:39 +00:00
c8245173a0
radv: pre-compute the line rasterization mode
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:39 +00:00
469350328c
radv: pre-compute the number of rasterization samples
...
The number of rasterization samples depend on many various states.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36988 >
2025-09-15 19:10:38 +00:00
8d991c2572
radv/meta: remove useless assertion when choosing resolve method
...
The destination image layout is used for depth/stencil resolves and
asserting isn't very useful.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:55 +00:00
c8f6b27964
radv/meta: simplify calling depth/stencil resolve helpers
...
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:55 +00:00
39725fc935
radv/meta: simplify barriers for resolves
...
This is equivalent.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:54 +00:00
e673ccfcb5
radv/meta: remove useless VK_ACCESS_2_SHADER_WRITE_BIT for subpass resolves
...
This doesn't do anything.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:54 +00:00
704fbbb108
radv/meta: rework depth/stencil resolves using graphics
...
This adds a new helper that doesn't depend on the rendering info.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:53 +00:00
141beaee4e
radv/meta: rework depth/stencil resolves using compute
...
This adds a new helper that doesn't depend on the rendering info.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:53 +00:00
2207d1e732
radv/meta: fix saving push constants for depth/stensil resolves on compute
...
Found by inspection.
Cc: mesa-stable
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37300 >
2025-09-15 18:52:52 +00:00
e3460f15fa
aco/opt: Work around GCC compiler issue
...
No functional change. Random code churn that, apparently, makes a
GCC miscompile disappear.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:21 +00:00
a06f38e5ae
aco/vn: Don't combine expressions across calls
...
This increases live state across calls, which in turn increases spilling
and makes for slower shaders overall.
On top of RT function calls:
Totals from 7 (0.01% of 81072) affected shaders:
Instrs: 8980 -> 8955 (-0.28%); split: -0.88%, +0.60%
CodeSize: 51976 -> 51684 (-0.56%); split: -1.02%, +0.46%
SpillSGPRs: 248 -> 244 (-1.61%); split: -3.63%, +2.02%
SpillVGPRs: 367 -> 365 (-0.54%); split: -1.09%, +0.54%
Scratch: 32768 -> 31744 (-3.12%)
Latency: 135669 -> 128720 (-5.12%); split: -5.13%, +0.01%
InvThroughput: 35301 -> 34783 (-1.47%); split: -1.51%, +0.05%
VClause: 241 -> 242 (+0.41%)
SClause: 117 -> 120 (+2.56%)
Copies: 1311 -> 1338 (+2.06%); split: -0.69%, +2.75%
PreSGPRs: 899 -> 895 (-0.44%); split: -1.56%, +1.11%
PreVGPRs: 1103 -> 1099 (-0.36%)
VALU: 6143 -> 6098 (-0.73%); split: -1.22%, +0.49%
SALU: 913 -> 933 (+2.19%); split: -0.11%, +2.30%
VMEM: 989 -> 967 (-2.22%)
SMEM: 201 -> 214 (+6.47%)
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:21 +00:00
575d3adbf5
aco/validate: Validate call instructions
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:21 +00:00
28dc185966
aco/sched: Handle calls
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
2be37a91fa
aco/live_var_analysis: Handle calls
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
9c8a17e172
aco/lower_to_hw_instr: Lower calls
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
3667a7b687
aco: Add call info
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
af812862b7
aco: Add call-related program/block properties
...
Indicates various properties about calls: Whether a program is an
indirect callee, whether a program or block contains function calls, and
whether registers used by a caller need to be preserved.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
917a98b722
aco: Add ABI and Pseudo CALL format
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
e850650f92
aco: Add function call attributes
...
ACO needs RADV to set certain attributes on NIR functions to help with
compilation of function calls.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
d18b438832
aco: Add RegisterDemand::operator!=
...
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/34531 >
2025-09-15 17:16:20 +00:00
e1ea3f8bbf
radv/video: Always use OBU_FRAME in AV1 encode
...
Saves couple bytes per frame.
Reviewed-by: Benjamin Cheng <benjamin.cheng@amd.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37176 >
2025-09-15 14:12:17 +00:00
310cfda034
ac/surface: add ac_compute_surface_modifier
...
Used by radeonsi to export existing texture modifier.
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com >
Reviewed-by: Marek Olšák <marek.olsak@amd.com >
Acked-by: Daniel Stone <daniels@collabora.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/31658 >
2025-09-15 09:39:19 +00:00
98cd68ec05
ac/surface: add radeonsi exported modifiers to supported list
...
radeonsi will export texture with these modifiers.
piglit tests:
spec@ext_image_dma_buf_import@ext_image_dma_buf_import-export-tex
spec@ext_image_dma_buf_import@ext_image_dma_buf_import-tex-modifier
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com >
Reviewed-by: Marek Olšák <marek.olsak@amd.com >
Acked-by: Daniel Stone <daniels@collabora.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/31658 >
2025-09-15 09:39:19 +00:00
1b0ec56c40
ac/surface: refine supported modifier list for multi block size
...
Reference KMD convert_tiling_flags_to_modifier(). And we are
going to add 4K block size.
Acked-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com >
Reviewed-by: Marek Olšák <marek.olsak@amd.com >
Acked-by: Daniel Stone <daniels@collabora.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/31658 >
2025-09-15 09:39:19 +00:00
5962659c85
radv: remove uses_rt from radv_shader_info
...
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:21 +00:00
a2d3cbac2a
radv: determine subgroup/wave size early
...
This means we can actually implement varying subgroup size correctly.
It also means that we implement the implicit SPIR-V 1.6 full subgroups
requirement in compute shaders with cswave32/rtwave32.
In the future it will also allow more optimizations that use the subgroup size.
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
The only somewhat complex case here is GFX10 geometry shaders, if gewave32 is
used. We then only know the subgroup size when is_ngg is decided, as legacy
GS doesn't support wave32.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:21 +00:00
76a502d75a
ac/nir: set subgroup size for gs copy shader
...
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:21 +00:00
c4cdbee2e6
radv: remove unused ballot_bit_size from shader info
...
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:21 +00:00
2cda56e8b7
ac/llvm: remove unused ballot size
...
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:20 +00:00
f83a6e6389
radv: add varying subgroup size to shader stage key
...
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37294 >
2025-09-14 13:21:20 +00:00
7db518cfe4
aco: Fixes warning: function get_branch_target/to_clrx_device_name defined but not used
...
../../src/amd/compiler/aco_print_asm.cpp:156:1: warning: 'bool aco::{anonymous}::get_branch_target(char**, aco::Program*, const std::vector<bool>&, char**)' defined but not used [-Wunused-function]
156 | get_branch_target(char** output, Program* program, const std::vector<bool>& referenced_blocks,
| ^~~~~~~~~~~~~~~~~
../../src/amd/compiler/aco_print_asm.cpp:105:1: warning: 'const char* aco::{anonymous}::to_clrx_device_name(amd_gfx_level, radeon_family)' defined but not used [-Wunused-function]
105 | to_clrx_device_name(amd_gfx_level gfx_level, radeon_family family)
| ^~~~~~~~~~~~~~~~~~~
Signed-off-by: Yonggang Luo <luoyonggang@gmail.com >
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37289 >
2025-09-13 08:23:07 +00:00
c183eb5bc8
radv: Flush L2 before CP DMA copy/fill when CP DMA doesn't use L2
...
In case the source or destination were previously written
through L2, we need to writeback L2 to avoid the CP DMA accessing
stale data.
However, as the CP DMA doesn't write L2 either, an invalidation
is also needed to make sure other clients don't access stale data
when they read it through L2 after the CP DMA is complete.
Doing an invalidation before the CP DMA operation should take
care of both.
Additionally, radv_src_access_flush also invalidates L2 before
the copied data can be read.
Cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36820 >
2025-09-13 05:18:28 +00:00
8ef2f492e2
radv: Add comment to document CP DMA prefetch
...
Explain which caches (MALL, L2) the prefetch works with.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36820 >
2025-09-13 05:18:28 +00:00
7f556805c1
radv: detect platform:virtio-mmio devices for virtgpu native context
...
VirtIO devices can be configured as platform devices instead of PCI,
which is especially common in microVM projects like libkrun.
Let's allow RADV to probe MMIO virtgpu devices.
Signed-off-by: Val Packett <val@invisiblethingslab.com >
Reviewed-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37281 >
2025-09-12 12:56:46 +00:00
4f94ca6c96
ac/llvm: fix get_global_address for global atomics
...
They do not have an ACCESS index.
Fixes: 9a33c03654konstantin.seurer@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37292 >
2025-09-12 08:55:39 +00:00
a52483d9e7
radv: fix capture/replay with sampler border color
...
The border color index must be captured and returned to the app,
otherwise it's broken if samplers aren't replayed in-order.
Cc: mesa-stable
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com >
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37291 >
2025-09-12 06:51:51 +00:00
Georg Lehmann