If both, any-hit and intersection shader, use scratch vars,
it could happen that they end up in the same location and
overwrite each other.
Found by inspection.
Fixes: c3d82a9622 ('radv: Add pass to lower anyhit shader into an intersection shader.')
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21863>
This should allow us to remove a big memset when compiling a
graphics pipeline. This is mostly for imported NIR stages which
don't go through radv_pipeline_stage_init().
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20947>
This should allow us to remove a big memset when compiling a
graphics pipeline. This is mostly for imported NIR stages which
don't go through radv_pipeline_stage_init().
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20947>
In file included from ../src/tool/pps/pps_device.cc:10:
../src/tool/pps/pps_device.h:23:11: error: ‘uint32_t’ does not name a type
23 | static uint32_t device_count();
| ^~~~~~~~
In file included from ../src/tool/pps/pps_counter.cc:10:
../src/tool/pps/pps_counter.h:22:4: error: ‘uint32_t’ does not name a type
22 | uint32_t id;
| ^~~~~~~~
Fixes: 1cc72b2aef ("pps: Gfx-pps v0.3.0")
Closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/8186
Signed-off-by: Vinson Lee <vlee@freedesktop.org>
Reviewed-by: Rob Clark <robclark@freedesktop.org>
Reviewed-by: David Heidelberg <david.heidelberg@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21714>
We have 3 new/changed functions with this commit:
1. _mesa_alloc_dispatch_tables creates all dispatch tables that are not
created on demand and sets them to nop. This operates on gl_dispatch,
so it's reusable (e.g. glthread will want to use it)
2. _mesa_free_dispatch_tables frees everything
3. _mesa_initialize_dispatch_tables initializes gl_dispatch for GL
(not glthread)
Acked-by: Adam Jackson <ajax@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21777>
I like this more. The name self-documents itself. It's always equal
to the dispatch set in glapi.
GLAPI is a definition, so can't use that.
Acked-by: Adam Jackson <ajax@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21777>
There is a new struct gl_dispatch, which I'd like to reuse in glthread.
This allows building code around gl_dispatch that can be shared between
mesa and glthread. This is only refactoring.
Acked-by: Adam Jackson <ajax@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21777>
Prepare for Clover removal; don't waste resources on Clover anymore.
Acked-by: Daniel Stone <daniels@collabora.com>
Reviewed-by: Eric Engestrom <eric@igalia.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Signed-off-by: David Heidelberg <david.heidelberg@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21865>
Upstream moved Host.h from Support to TargetParser in LLVM 17.
This shouldn't lead to a FTBFS, since there is a forwarding include left
behind. Sadly the added deprecation warning #pragma is invalid and thus
causes a build failure right away. But since we would have to follow the
move anyway in the future, just do it right away.
Reference: d768bf994f
Reviewed-by: Bas Nieuwenhuizen <bas@basnieuwenhuizen.nl>
Closes: #8275
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21263>
After enabling the Wayland platform for x86_64,
multiple new tests were triggered, some of which timed out.
Also wayland-dEQP-EGL.functional.negative_api.create_pixmap_surface now pass.
Signed-off-by: David Heidelberg <david.heidelberg@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21786>
Currently, the generated tests consist of some boilerplate, generated
test cases, and at the very end the actual test. This is bad for readability,
because the actual code is all the way at the bottom. It's also bad for
clang-format linting: even though the test cases are /* clang-format off */,
they still take an exceptionally long time to parse when linting. I suspect this
is a clang-format bug, but it's easy enough to workaround.
To solve these issues, restructure so that the test cases are in separate files
(containing the actual data), but the manually written test functions are
consolidated into a new family of generated layout tests. This is probably
cleaner.
Parallel clang-format linting is now 10x faster on the M1, which means it's
now practical to lint in my "publish branch" hook.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21854>
Not using it yet, that will be done in the next patch.
Xe only supports submission using VM.
For i915 the backend functions are just a noop.
Signed-off-by: José Roberto de Souza <jose.souza@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21698>
As noted on [1], lowering patterns of the form
floatS -> floatB -> floatS ==> floatS
cannot require precision since this may cause flush denorming.
[1] 3f779013 ("nir: Add an algebraic optimization for float->double->float")
Fixes: b86305bb ("nir/algebraic: collapse conversion opcodes (many patterns)")
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Signed-off-by: Isabella Basso <isabellabdoamaral@usp.br>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20965>
Conversions on smaller bit sizes should also be collapsed when composed.
This also adds more patterns on the
intS -> intB -> floatB ==> intS -> floatB
lowering so as to deal with any int size C > B instead of a fixed intB.
Closes: #7776
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Signed-off-by: Isabella Basso <isabellabdoamaral@usp.br>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20965>
Some patterns were outside the list of optimizations.
Fixes: b86305bb ("nir/algebraic: collapse conversion opcodes (many patterns)")
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Signed-off-by: Isabella Basso <isabellabdoamaral@usp.br>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20965>
Rather than ingesting separate control and memory barriers, ingest only the
combined and optimized scoped_barrier intrinsic. For barriers originating from
GLSL, this makes it easier to ensure correctness. For barriers originating from
SPIR-V, this is required for translation at all, as spirv_to_nir knows only
scoped barriers. So this gets us closer to Vulkan and OpenCL.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21752>
New 10 devices - asus-CM1400CXA-dalboz hosted on Collabora farm.
1x Move VA-API tests to the dalboz (more resources). One timeout dropped.
9x Run VKCTS on dalboz.
Signed-off-by: David Heidelberg <david.heidelberg@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21702>
Lowers away 64-bit loads, which we'll create in the sysval lowering for
dynamically indexed UBOs/VBOs. The lowering generates pack_64_2x32 instructions,
so lower those too.
No shader-db changes.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21674>
Instead of lowering to bitwise ops. Yet another way of subdividing in NIR.
Probably insignificant but makes it easy to check that the pass ordering from the
previous pass is right. It does let us get much better codegen for
unpacksnorm2x16, whatever that's worth.
No shader-db changes.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21674>
As intended. We can't CSE with partial null destinations in the way, so we
shouldn't eliminate dead destinations until after CSE has run. But we should
still eliminate dead instructions to ensure CSE doesn't move things around
needlessly, hurting register pressure.
Noticed while debugging live range splitting.
No GLES3.0 shader-db changes.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21674>
Our dead code elimination pass does two things:
1. delete instructions that are entirely unnecessary
2. delete unnecessary destinations of necessary instructions
To deal with pass ordering issues, we sometimes want to do #1 without #2.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21674>
We should handle nir_op_unpack_32_2x16_split_* natively, since we can generate
better code with agx_subdivide (coalescing the ops away) than the bitshift
lowering.
That said, we do need some extra instructions for the floating point
conversions.
No shader-db changes (which makes sense because we're targetting the GLES3.0
shader-db, which doesn't have the packing GLSL functions).
The real motivation of this change isn't optimizing some GLSL pack functions,
though, it's avoiding a code regression from using NIR's memory bit size
lowering in a future MR. That lowering will turn things like "load i16vec4" into
"load i32vec2 + unpack_32_2x16", so we need to be able to coalesce that unpack.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21674>
At some point in a refactoring long ago, our 'Piglit' runs on arm64
started actually being dEQP-GLES2 runs. Oh dear.
Surprisingly, there are a number of expectation changes; added every
fail I saw from a long overnight stress test.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21851>
Minimum/maximum LOD and LOD bias are unsigned and signed fixed point formats
respectively. They are not unsigned integers. Introduce fixed-point types into
our GenXML and use them in the XML, rather than packing in sidebands. This makes
the XML more correct and fixes pretty-printing of texture and sampler
descriptors.
Signed-off-by: Alyssa Rosenzweig <alyssa@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20445>
Connor Abbott wrote a nice explanation of how instance divisors work on Mali.
Let's add it to the driver docs instead of letting it languish in a forgotten
header file.
This is mostly pasted from the existing header in tree, with a few local changes
applied.
Signed-off-by: Alyssa Rosenzweig <alyssa@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20445>
Nowadays, formats are defined with GenXML, not the old panfrost-job.h, so most
of the format #defines in panfrost-job.h are unused. That said, a few are still
in use as a backdoor for compressed format queries to avoid a GenXML dependency.
That's not great but cleaning that up isn't the subject of this MR.
Signed-off-by: Alyssa Rosenzweig <alyssa@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20445>
XE architecture enables many more metrics, perhaps too many for
the average user. Reduce reported metrics to smaller subset,
known as non-extended metrics, by default. Can re-enable extended
metrics with env var INTEL_EXTENDED_METRICS=1
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21841>
Implement buffer textures in full generality. There are a few issues here:
* OpenGL requires buffer textures support a minimum size of 65536 elements,
however 1D textures in AGX are (at most) 8192 elements.
* OpenGL 4.0 (and OpenGL ES) require buffer textures to support the "RGB32"
texture formats. These are 3 packed channels of 32-bits each. In general,
non-power-of-two texel sizes are problematic. AGX does not support any such
formats and we rely on the GL frontend to lower to a padded format (RGBX) if
necessary. Such a lowering cannot work for buffer textures, however, so we
need to find a way to implement RGB32 buffer textures.
We solve these issues in the follow way:
* Use 2D texture descriptors for buffer textures, with a large fixed
power-of-two size along one axis. Then large texel indices may be accessed at
a small vec2 texel coordinate, and since the fixed dimension is a
power-of-two, that vector may be recovered by simply shifting and masking.
This effectively avoids size restriction. We do need to clamp texel indices to
the buffer size to avoid faulting on OOB reads, since we may read past the end
of the buffer (if the app binds a non-page-aligned offset into the buffer).
* Use a general purpose memory load for RGB32 buffer textures. Lower the texture
load instruction to a memory load from the buffer and some address arithmetic.
There's no format conversion needed for RGB32, other than maybe filling in a
format-appropriate alpha, so this is straightforward. Again, we need to clamp
the texel index for robustness with OOB reads.
Each of these solutions brings its own problem.
* Using 2D textures instead of 1D requires physically rounding up the buffer
size when packing the descriptor, so we can no longer implement textureSize()
by reading off the texture descriptor like normal.
* We don't know at compile-time whether a given texture load will read from an
RGB32 buffer texture or not, so we need to emit code for both. In Vulkan, we
can't key the shader to this property, either, since it's descriptor set state
and not pipeline state.
And each of these problems in turn brings its own solution:
* The texture descriptor is linear, so the "compression buffer address" field is
ignored by the hardware. We stash the real buffer size there so that
textureSize becomes a load from the texture descriptor like usual, without
requiring a sideband (which would complicate bindless textures).
* If we determine a texture descriptor contains RGB32 data, then it will never
be interpreted by the hardware and hence does not need to be a valid texture
descriptor. So, we extend the hardware's format enum to contain a
software-defined RGB32 format enum. Then, when lowering texture buffer loads,
we either read it as a typed RGB32 memory load or as a texture load depending
on the value of the format field in the texture descriptor.
All of this is accomplished with a big NIR pass generating a pile of strange
looking code. But it should be good enough in practice for this silly feature.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21672>
In NIR, texelFetch (txf) does not use a sampler, but in AGX, it does -- even
though the contents of the sampler are semantically irrelevant. Rather than
requiring the state tracker to bind a sampler anyway (indicated for texture
buffers with PIPE_CAP_TEXTURE_BUFFER_SAMPLER), just add a dummy sampler
ourselves if txf is used and there are otherwise no samplers. This is helpful
because PIPE_CAP_TEXTURE_BUFFER_SAMPLER isn't honoured by Rusticl or seemingly
mesa/st's PBO code, and after implementing this dummy sampler workaround in
Panfrost for Rusticl, I realized this CAP is silly and shouldn't exist in the
first place. (And I regret pushing for its reinclusion.)
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21672>
This commit ensures that we are using mesa release builds in performance
jobs.
To achieve that, some modifications were made on top of
https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21492.
- Append the `BUILDTYPE` variable into the S3 artifact name
(MINIO_ARTIFACT_NAME environment variable) to allow for better
artifact management.
- The ./artifacts directory has been added to the list of artifact
directories for build-common. This ensures that the debian-release and
debian-arm64-release jobs are the only ones necessary for running
performance jobs. These jobs only produce artifacts via
prepare-artifacts.sh when we are under performance workflow.
- Make lava-submit.sh behave similar to baremetal jobs regarding
MINIO_ARTIFACT_NAME variable. For example, users can now easily
differentiate between mesa-arm64.tar.zstd and
mesa-arm64-release.tar.zstd by looking inside the `Downloading
artifacts from s3` Gitlab section.
Signed-off-by: Guilherme Gallo <guilherme.gallo@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21804>
The check in alloc_bo_from_cache() was skiping any try to get a bo
from cache but after use a protected bo was still being put in some
cache bucket and could be used for cases that don't require a
protected bo.
Using a protected bo in cases that don't require it can have
performance implications.
So here returning NULL when trying to get a cache bucket for a
protected bo, this will cause bo->real.reusable to be set to false
avoiding the bo to be reused.
Fixes: 9402ac8023 ("iris: handle protected BO creation")
Signed-off-by: José Roberto de Souza <jose.souza@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21824>
We don't have a way to tell the ZLS hardware to use linear buffers, so if a
buffer could be used for depth/stencil, we have to twiddle. This isn't a problem
in practice, since depth/stencil buffers can't be shared across processes or
mapped directly as linear.
Fixes faults in depthstencil-render-miplevels, which was picking linear for one
buffer because of a STAGING bind flag. But that won't work :-)
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21753>
On Intel platforms, the uclz lowering if ufind_msb is either one
instruction better (Gfx7 and newer) or two instructions better (all
older platforms) than the ifind_msb implementations.
On platforms that use lower_find_msb_to_reverse, there should be no
difference.
All Haswell and newer Intel platforms had similar results. (Ice Lake shown)
total instructions in shared programs: 19938662 -> 19938634 (<.01%)
instructions in affected programs: 850 -> 822 (-3.29%)
helped: 2 / HURT: 0
total cycles in shared programs: 858467067 -> 858465538 (<.01%)
cycles in affected programs: 10080 -> 8551 (-15.17%)
helped: 2 / HURT: 0
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19042>
4d802df3aa loosened the type restrictions
on these opcodes to enable support for 64-bit ballot operations. In
doing so, it enabled 8-bit and 16-bit sizes as well.
It's impossible to get these sizes through GLSL or SPIR-V. None of the
lowering in nir_opt_algebraic can handle non-32-bit sizes. Almost no
drivers can handle non-32-bit sizes.
It doesn't seem possible to enforce anything other than "one bit size"
or "all bit sizes" in nir_opcodes.py. The only way it seems possible to
enforce this is in nir_validate. This is not ideal, but it be what it
be.
v2: Remove restriction on find_lsb. It is acutally possible to get this
via GLSL by doing findLSB() on a lowp value. findMSB declares its
parameter as highp, so that path is still impossible.
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19042>
Fossil-db results:
All Intel platforms had similar results. (Ice Lake shown)
Cycles in all programs: 9098346105 -> 9098333765 (-0.0%)
Cycles helped: 6
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19042>
The 31-ufind_msb_rev(x) lowering only produces the correct result for
32-bit sources. ufind_msb_rev can also have 64-bit sources, and most
platforms are expected to lower this to 32-bit instructions with extra
logic operations.
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19042>
Unlike ufind_msb, ifind_msb is only defined in NIR for 32-bit values, so
no @32 annotation is required.
No shader-db or fossil-db changes on any Intel platform.
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19042>
Stipple lines now appear correctly when they are oblique.
Previously the number of steps of the stipple counter between two vertices
was calculated as the euclidian distance between them in screen space, however
the length occupied by pixel along a line is only `1` for lines that are either
vertical or horizontal and will be anywhere between `1` and `sqrt(2)`
for other cases.
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21290>
Get the texture/sampler index from the texture/sampler_offset source (which
is an offset from 0 thanks to the lower_index_to_offset lowering) and feed it in
as corresponding 16-bit texture instruction sources.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21704>
For indirect indexing into the binding table. Note this does not handle packing
the bindless forms, since that's a bit more involved.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21704>
This is an inline function with a compile-constant switch, so I expect
the compiler wouldn't produce any better code like this, but for humans
it's easier to read when function calls are not embedded into other
function calls.
Signed-off-by: Eric Engestrom <eric@igalia.com>
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21835>
When enabled, on gfx12 plus, we will add the sync nop instruction after
each instruction to make sure that current instruction depends on the
previous instruction explicitly.
This option will help us to get a hint if something is missing or broken
in software scoreboard pass.
Signed-off-by: Sagar Ghuge <sagar.ghuge@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21797>
Now that we aren't using them on Gfx8+ we can drop a lot of cruft.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
This mainly lets the software scoreboarding pass correctly mark the
instructions, without needing to resort to fragile manual handling in
the generator.
We can also make small improvements. On Gfx 8LP-12.0, we no longer have
the restrictions about DWord alignment, so we can simply write each half
into its intended location, rather than writing it to the low DWord and
then shifting it in place.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
I originally thought that we were intentionally emitting the legacy
opcodes here to make them opaque to the optimizer, so that it wouldn't
eliminate the explicit type conversions, as they're actually required
to do the quantization. But...we don't actually optimize those away
currently anyway. So...go ahead and use the helpers for consistency.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
With the previous patch, we no longer need to special case this, as we
emit a MOV with an HF source, rather than F16TO32 with an UW source,
on all platforms that need scoreboarding.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
This gets us a MOV at the IR level on Gfx8+ which should be more
optimizable than F16TO32. It also removes confusion about which
pipe which the instruction will run on.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
We can just use the new builder helpers to get the optimization
advantages of a MOV on Gfx8+ while also getting the necessary F32TO16
on Gfx7.x and yet not worry too hard about it.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
These take care of emitting the F32TO16/F16TO32 instructions on Gfx7.x
but otherwise just emit a type converting MOV on Gfx8+.
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
For converting half-float to float, we currently emit BRW_OPCODE_F16TO32
with a UW source, to match legacy Gfx7 behavior. In the generator, this
becomes a MOV with a HF source on Gfx8+. Unfortunately, this UW source
confuses the scoreboarding pass into thinking it's an integer source,
leading to incorrect SWSB annotations on Alchemist.
We should ultimately fix the IR to stop being so...legacy...here, but
this is the simplest fix for stable branches.
Fixes misrendering in Elden Ring and likely Sekiro: Shadows Die Twice.
Cc: mesa-stable
Tested-by: Chuansheng Liu <chuansheng.liu@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Sagar Ghuge <sagar.ghuge@intel.com>
References: https://gitlab.freedesktop.org/mesa/mesa/-/issues/8018
References: https://gitlab.freedesktop.org/mesa/mesa/-/issues/8375
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21783>
Wa_1409600907 was enabled for gen12+. It should not be applied for
platforms after gen12.0. Use generated helpers to ensure application
to all relevant platforms.
Reviewed-by: Tapani Pälli <tapani.palli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21743>
We used to set RADEON_FLAG_GTT_WC when wsi_info is set. This changes it
to set the flag for any external mem on vram, extending the logic for
apps using external memory directly.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21803>
CTS runs on stoney are currently taking ~20min to complete, which seems
to have begun with the upgrade to CTS 1.3.5.0. This is a bit too long in
and of itself, but it means that - assuming zero contention - a job that
has to be retried because the machine hung can take 40 minutes.
Aim to drop this to 15min turnaround by lowering the overall fraction
from 1/8th of the CTS to 1/11th.
As the jobs we run have been reshuffled, this adds a lot more expected
fails. As most of them categorise easily into patterns, group the
failures together in the file. Non-strict wide lines has passed since we
last ran it; the other failures all group into existing classes seen
for a long time.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21791>
Quake II uses GL_POINT_SMOOTH to render particles.
Zink currently requires `zink_emulate_point_smooth` to support that feature.
Acked-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21731>
The HW can technically execute 16-bit operations, but the restrictions on
16-bit ALU ops are so great that it ends up not being a win for
GLES-on-Vulkan to lower mediump to 16-bit operations, at least with the
current state of the Intel compiler. This brings zink-on-anv in line with
iris and angle-on-anv for mediump behavior (ANGLE uses RelaxedPrecision,
which we ignore).
Perf on some angle traces on my brya (ADL) and i9-9900K (CFL):
ADL zink pubg_mobile_battle_royale: +13.4574% +/- 5.2046% (n=5)
CFL zink pubg_mobile_battle_royale: +29.5332% +/- 0.646585% (n=6)
ADL zink aztec_ruins_high: +5.78027% +/- 4.80645% (n=4)
CFL zink aztec_ruins_high: -1.10641% +/- 0.140562% (n=12)
ADL zink trex_200: +5.86956% +/- 2.09633% (n=10)
CFL zink trex_200: +9.72136% +/- 0.749261% (n=10)
Reviewed-by: Ivan Briano <ivan.briano@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21775>
Our TGL machines are currently slightly oversubscribed (max. 17 jobs in
a pipeline on 15 DUTs). They're also currently suffering from
thermally-induced GPU throttling (being investigated), and a
thundering-herd network load effect: as all 15 jobs start at once, we
end up saturating one of our network links.
The combination of all three of these things means that TGL is often our
long pole in CI runs. Until we can ameliorate the two issues
constraining throughput (and a third where an unreliable hardware UART
sometimes kills jobs when it shouldn't), halve the workload so we at
least have some breathing room to absorb them.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21790>
Seems to fix a hang in the following titles :
- Age of Empire 4
- Monster Hunter Rise
where the HW is hung on a PIPE_CONTROL after a GPGPU_WALKER but no
MEDIA_INTERFACE_DESCRIPTOR_LOAD was emitted since the switch from 3D
to GPGPU.
This would happen in the following case :
vkCmdBindPipeline(COMPUTE, cs_pipeline);
vkCmdDispatch(...);
vkCmdBindPipeline(GRAPHICS, gfx_pipeline);
vkCmdDraw(...);
vkCmdDispatch(...);
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Ivan Briano <ivan.briano@intel.com>
Cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/17247>
The previous solution breaks potential loop header phis.
Move the dummy-break to the bottom of the loop.
Fixes: dEQP-VK.reconvergence.subgroup_uniform_control_flow_ballot.*
Fixes: a9c4a31d8d ('aco: handle NIR loops without breaks')
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21736>
the recording rp_info may be a pointer to a member of the array being
reallocated, so test for this and re-set it to avoid invalid memory
access
found with this caselist:
KHR-GL46.texture_gather.offset-gather-unorm-2darray
KHR-GL46.texture_view.view_sampling
cc: mesa-stable
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21729>
PA_SC_VRS_OVERRIDE_CNTL is emitted when a framebuffer is bound because
it controls the VRS surface enable bit. Though, if a pipeline is bound
after the framebuffer is emitted, it can override the state. Remove it
completely since VRS for flat shading and RADV_FORCE_VRS are disabled.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20333>
On GFX11, VRS rate images can't use linear tiling and the swizzle mode
must be either SW_Z or SW_R.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20333>
NGG streamout performance is limited by the workgroup size, so make it as
large as possible.
Since this uses si_get_max_workgroup_size() to set the NGG workgroup size,
the side effect is that all GS is also getting an increase to 256, which
is OK.
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21403>
This should help with NGG streamout performance, which is limited by
the workgroup size (it should be as large as possible).
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21403>
This moves the code to a separate branch to make it less intertwined
with the rest to allow sampler descriptor lowering later.
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Qiang Yu <yuq825@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21403>
The hardware uses the register to premultiply GS vertex indices
in input VGPRs.
This changes the behavior as follows:
- VGT_ESGS_RING_ITEMSIZE is always 1 on gfx9-11, set in the preamble.
- The value is passed to the shader via current_gs_state (vs_state_bits).
- The shader does the multiplication.
The reason is that VGT_ESGS_RING_ITEMSIZE will be removed in the future.
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21403>
The naming of aco_*_key didn't make sense because they
were never actually used as cache keys, only radv_*_key
are used as cache keys.
Rename the aco structs to aco_*_info instead.
Signed-off-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21696>
v2 by Timur Kristóf:
- Rebase this patch on latest main.
Signed-off-by: Qiang Yu <yuq825@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21696>
v2 by Timur Kristóf:
- Rebase this patch.
Signed-off-by: Qiang Yu <yuq825@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21696>
Handle arguments that need a waitcnt without relying on
RADV specific VS input information.
Signed-off-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21696>
This is to indicate when an argument was loaded from VMEM
and needs a waitcnt before it can be used.
Signed-off-by: Timur Kristóf <timur.kristof@gmail.com>
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21696>
Sometimes we can end up with a sequence where we need to flush a batch
with no clears and no draws (for ex, to get a fence). Promote these to
sysmem.
Signed-off-by: Rob Clark <robdclark@chromium.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21747>
This patch will add support for dpb resize when low to high resolution
change/ svc use-cases.
With DPB tier1 type,vp9 svc decoder use cases are failed. This
Change will fix this[VCN1/VCN2].
Signed-off-by: SureshGuttula <suresh.guttula@amd.com>
Reviewed-by: Leo Liu <leo.liu@amd.com>
Reviewed-by: Boyuan Zhang <boyuan.zhang@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21548>
The gallium disk cache is about to depend on these, and I don't want to
create a dependency on agx_opcodes.h.py for that. So, make a new header
for them that doesn't have build dependencies.
Rename them to agx_compiler_* too, to avoid collisions with the other
driver debug flags.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21776>
The function _mesa_symbol_table_add_symbol() copies the string with strdup(),
the original string should be freed.
For instance, with "piglit/fp-fragment-position -auto -fbo":
Direct leak of 7 byte(s) in 1 object(s) allocated from:
#0 0xffff99c59050 in __interceptor_strdup (/usr/lib64/libasan.so.6+0x59050)
#1 0xffff8f53d24c in handle_ident ../src/mesa/program/program_lexer.l:129
#2 0xffff8f53d24c in _mesa_program_lexer_lex ../src/mesa/program/program_lexer.l:312
#3 0xffff8f529d10 in yylex ../src/mesa/program/program_parse.y:289
#4 0xffff8f529d10 in yyparse src/mesa/program/program_parse.tab.c:2140
#5 0xffff8f5341a4 in _mesa_parse_arb_program ../src/mesa/program/program_parse.y:2589
#6 0xffff8f51e96c in _mesa_parse_arb_fragment_program ../src/mesa/program/arbprogparse.c:82
#7 0xffff8f4d867c in set_program_string ../src/mesa/main/arbprogram.c:402
Signed-off-by: Patrick Lerda <patrick9876@free.fr>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21728>
For instance, with "piglit-2000/bin/asmparsertest ARBfp1.0 tests/asmparsertest/shaders/ARBfp1.0/shadow-02.txt":
Direct leak of 192 byte(s) in 2 object(s) allocated from:
#0 0x7f6e8378f987 in calloc (/usr/lib64/libasan.so.6+0xb1987)
#1 0x7f6e7769d620 in asm_instruction_copy_ctor ../src/mesa/program/program_parse.y:2146
#2 0x7f6e7769d620 in yyparse ../src/mesa/program/program_parse.y:439
#3 0x7f6e776a6725 in _mesa_parse_arb_program ../src/mesa/program/program_parse.y:2590
#4 0x7f6e77687f69 in _mesa_parse_arb_fragment_program ../src/mesa/program/arbprogparse.c:82
#5 0x7f6e77630765 in set_program_string ../src/mesa/main/arbprogram.c:402
#6 0x7f6e76ec3e8a in _mesa_unmarshal_ProgramStringARB src/mapi/glapi/gen/marshal_generated2.c:4152
#7 0x7f6e76a0e585 in glthread_unmarshal_batch ../src/mesa/main/glthread.c:122
#8 0x7f6e76a1031d in _mesa_glthread_finish ../src/mesa/main/glthread.c:383
#9 0x7f6e76a1031d in _mesa_glthread_finish ../src/mesa/main/glthread.c:348
#10 0x7f6e76e6a062 in _mesa_marshal_GetError src/mapi/glapi/gen/marshal_generated1.c:1809
Signed-off-by: Patrick Lerda <patrick9876@free.fr>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21728>
For instance, with "piglit/bin/asmparsertest ARBfp1.0 tests/asmparsertest/shaders/ARBfp1.0/swz-04.txt":
Direct leak of 18 byte(s) in 2 object(s) allocated from:
#0 0x7f97e99050 in __interceptor_strdup (/usr/lib64/libasan.so.6+0x59050)
#1 0x7f8d4160ac in handle_ident ../src/mesa/program/program_lexer.l:129
#2 0x7f8d4160ac in _mesa_program_lexer_lex ../src/mesa/program/program_lexer.l:312
#3 0x7f8d402b50 in yylex ../src/mesa/program/program_parse.y:289
#4 0x7f8d402b50 in yyparse src/mesa/program/program_parse.tab.c:2140
#5 0x7f8d40d01c in _mesa_parse_arb_program ../src/mesa/program/program_parse.y:2590
#6 0x7f8d3f77ac in _mesa_parse_arb_fragment_program ../src/mesa/program/arbprogparse.c:82
#7 0x7f8d3ad468 in set_program_string ../src/mesa/main/arbprogram.c:402
Signed-off-by: Patrick Lerda <patrick9876@free.fr>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21728>
This is required by register shadowing (required by the new PAIRS packets),
preemption, user queues, and we only have to wait for VS after streamout,
not PS. This is how gfx11 streamout should have been done.
Reviewed-by: Qiang Yu <yuq825@gmail.com>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21584>
Wa_14017989577 is a clone of Wa_14015360517, which applies to several
platforms beyond INTEL_PLATFORM_DG2_G10.
Update references to Wa_14017989577, and use the generated workaround
helper to ensure application to the proper platforms.
Reviewed-by: Tapani Pälli <tapani.palli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21744>
Both OpenGL and Vulkan drivers share the same performance counters.
Let's move them to a common place instead of duplicating.
Signed-off-by: Juan A. Suarez Romero <jasuarez@igalia.com>
Reviewed-by: Iago Toral Quiroga <itoral@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21420>
For the platforms that call it, it a function in the hot path so
moving it to kmd backend.
After this patch i915/iris_bufmgr.c is empty but not removing it
as next patch will add functions to it.
Signed-off-by: José Roberto de Souza <jose.souza@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21494>
bo_madvise() is on hot path, so moving it to kmd backend.
Signed-off-by: José Roberto de Souza <jose.souza@intel.com>
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21494>
Swizzle of 8-bit stencil format is defined as _x__ but the hw expects
BC_SWIZZLE_XYZW.
Fixes dEQP-VK.pipeline.monolithic.sampler.border_swizzle.*s8_uint*.
Cc: mesa-stable
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21482>
Modifying pitch for all LINEAR surface isn't correct;
the original change that modified surf_pitch was only
intended for YUV textures.
This fixes vkGetImageSubresourceLayout rowPitch return value
for VK_FORMAT_BC3_UNORM_BLOCK + VK_IMAGE_TILING_LINEAR.
Fixes: fcc499d5 (ac/surface: adjust gfx9.pitch[*] based on surf->blk_w)
v2: add check for UYVY format (Pierre-Eric)
v3: move blk_w division to above if check (Pierre-Eric)
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Reviewed-by: Bas Nieuwenhuizen <bas@basnieuwenhuizen.nl>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21595>
It's just v_fma with fixed DPP8 and builtin s_waitcnt_expcnt, so it can mostly
be handled as a pure VALU instruction.
Reviewed-by: Daniel Schürmann <daniel@schuermann.dev>
Reviewed-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21023>
Both GLSL & SPIRV have undefined values for shift > bitsize. But SM5
says :
"This instruction performs a component-wise shift of each 32-bit
value in src0 left by an unsigned integer bit count provided by
the LSB 5 bits (0-31 range) in src1, inserting 0."
Better to not hard code the wrong behavior in NIR.
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Fixes: e227bb9fd5 ("nir/builder: add ishl_imm helper")
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@colllabora.com>
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21720>
It's unnecessary and already checked elsewhere like in
vkCmdBindDescriptorSets(). This improves performance of vkoverhead
test #76 (descriptor_1image) by +18%. It's the same performance as
PRO on my Threadripper 1950X now. This should also slightly improve
texel and buffer descriptors.
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20699>
We can't add a 64-bit immediate with the hardware iadd, that won't work. What we
can do is add a 32-bit immediate, derived as the low 32-bits of a 64-bit
nir_ssa_def.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21643>
If we manage to fold in a left shift that's bigger than the hardware can do, we
should at least avoid generating a useless right shift to feed the hardware
rather bailing completely.
For motivation, this form of address arithmetic is encountered when indexing
into arrays with large power-of-two element sizes (array-of-structs).
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21643>
Optimize address arithmetic of the form
base + u2u64((index << shift) + const)
into hardware operands
base, index << (shift - format_shift) + const'
which (if format_shift = shift) can be simply
base, index + const'
rather than the current naive translation
base, ((index << shift) + const) >> format_shift
This saves at least one pointless shift. We can't do this optimization with
nir_opt_algebraic, because explicitly optimizing "(a << #b) >> #b" to "a" isn't
sound due to overflow. But there's no overflow issue here, which is what this
whole pass is designed around.
For motivation, this address arithmetic implements "dynamically indexing into an
array inside of a C structure", where the const is the offset of the array
relative to the structure.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21643>
Once we've matched a summand, commit to it. This avoids needlessly checking the
second source if the first matched, and removes some indentation/funny control
flow.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21643>
this otherwise may have been a surface that was never drawn to or
already had its layout corrected, in which case a deferred barrier
is not only unnecessary, it might be broken
cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21739>
Wa_1209978020 is a clone of Wa_18012201914. Update references to
refer to the originating bug, and use generated helpers to ensure it
is applied to future platforms as needed.
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21741>
Rather than control_barrier. This avoids the need to handle control_barrier at
all for backends that set use_scoped_barrier. This effectively matches what
spirv_to_nir emits, so Vulkan-capable compilers should be ok.
Signed-off-by: Alyssa Rosenzweig <alyssa@collabora.com>
Reviewed-by: Emma Anholt <emma@anholt.net>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21634>
Now that we handle scoped barriers with execution scope during
NIR -> Backend IR translation, this lowering is not needed anymore.
Reviewed-by: Alyssa Rosenzweig <alyssa@collabora.com>
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21634>
For GLSL, we want to optimize code like
memoryBarrierBuffer();
controlBarrier();
into a single scoped_barrier intrinsic for the backend to consume. Now that
backends can get scoped_barriers everywhere, what's left is enabling backends to
combine these barriers together. We already have an Intel-specific pass for
combining memory barriers; it just needs a teensy bit of generalization to allow
combining all sorts of barriers together.
This avoids code quality regression on Asahi when switching to purely scoped
barriers. It's probably useful for other backends too.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21661>
Some backends can handle a constant texture index or a dynamic texture index but
not a constant texture index plus a dynamic texture offset. Add a nir_lower_tex
option to lower to one of these options.
v2: Use more straightforward code proposed by Faith.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Emma Anholt <emma@anholt.net> [v1]
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21546>
This may happen if the Nouveau device fails to initialize before
the pushbuf is set up; the abort logic will call this either way.
Signed-off-by: Sam Edwards <CFSworks@gmail.com>
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21611>
Though not many people run full test runs, it occupies 2/7 a630 slots
for nearly 2 hours. If more than one person does this at a time, it can
be an effective DoS and make merges time out.
Limit full runs to a subset of the runners, such that at least some of
them will always be available for us.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21737>
this ensures that batch refs are added for fb surfaces on unbind, which
prevents stale batch tracking from persisting on resources
after the context is destroyed
fixes:
*EGL.functional.render.multi_context*
cc: mesa-stable
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21727>
GALLIUM_HUD_OPACITY can be used with a % value from 0 to 100, to specify
the opacity of the gallium hud background.
Signed-off-by: Italo Nicola <italonicola@collabora.com>
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Gert Wollny <gert.wollny@collabora.com>
Tested-by: Chris Healy <healych@amazon.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/20339>
reason:
some h264 streams have some strange pictures, from
vaapi input these pictures don't have a reference frame,
however, they are not intra only pictures, in MB layer
these pictures are looking for some references, if they
cannot find it. It could cause PF.
when reference pictures exist, it will need to set used_for
reference_flags, therefore if that is set, however the
number of reference frames is zero, which is odd, it
should be avoided.
solution:
In the above case, to scan the ref list so that it will
make at least one reference available to avoid crash, since
this is not accurate enough, it could cause some artifacts.
And in that case, it will need to be checked individually
for another solution.
closes: https://gitlab.freedesktop.org/drm/amd/-/issues/1462
closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/8401
Cc: mesa-stable
Tested-by: llyyr <llyyr.public@gmail.com>
Reviewed-by: Leo Liu <leo.liu@amd.com>
Signed-off-by: Ruijing Dong <ruijing.dong@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21732>
8bit av1 decoding cannot use NV12 output, which is caused by missing NV12 RT format.
It's a regression caused by commit [0504a3b0].
Fixes: 0504a3b0 ("radeonsi: correctly declare YUV420_10 RT Format support for AV1")
Cc: mesa-stable
Signed-off-by: Sonny Jiang <sonny.jiang@amd.com>
Reviewed-by: David Wu <David.Wu3@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21697>
Hit this with vulkaninfo on X11, probably a vulkaninfo bug. Though
moving the assertion doesn't hurt.
Fixes: be0dcbdfa2 ("wsi/x11: Implement EXT_swapchain_maintenance1.")
Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Reviewed-by: Hans-Kristian Arntzen <post@arntzen-software.no>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21722>
these don't actually work; it creates the marker for the op, but
then the "end" of the marker is effectively determined to be the end
of the cmdbuf
instead, detect whether a draw is from u_blitter and add a marker there
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21583>
This is preliminary work for separate shader functions.
The ray_tracing_module is eventually intended as self-contained
pipeline struct per RT group.
For now, these modules only contain the group handles.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21667>
It can sometimes be useful to also print the shaders that are marked as
internal, so let's add a flag that lets us do that.
Reviewed-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21681>
* if the proposed box is smaller than an existing box then don't add it,
* if the proposed box is adjacent to an existing box, expand
* if the proposed box is larger than an existing box, replace
this reduces the chances of having a ton of copy boxes to iterate over
also add a perf warning in case a ton of copy boxes exist
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21665>
if a single query is being started and stopped frequently, update
the internal qbo with a single copy call instead of one copy per result
not actually that useful in practice because of how query pools are shared,
but could help somewhere in theory
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21581>
for drivers lacking one of:
* EXT_color_write_enable
* primitivesGeneratedQueryWithRasterizerDiscard
terrible things must happen. specifically, dummy surfaces have to
be used in a framebuffer with rast-discard enabled for the duration
of the query
now that queries are only started/stopped in renderpasses, however,
there are new hurdles. with tc renderpass optimizing, queries can be
started outside the renderpass, which would trigger recursion when
trying to start a primgen query outside the renderpass if any clears
are enabled, as those must be flushed onto the real surfaces
to solve all of this:
* block tc renderpass optimizing if at least one of the above features is missing
* detect a pending primgen query start during renderpass start
* activate rast-discard and set dummy surfaces before beginning renderpass
* this recurses and automatically flushes clears
* finally, start the real renderpass
BUT WAIT THERE'S MORE!
because there's also drivers that support EXT_color_write_enable and don't support
primitivesGeneratedQueryWithRasterizerDiscard, which means they do need rast-discard,
but they don't need dummy surfaces, and so the clears still have to be flushed,
so they need an explicit (recursive) renderpass start/stop in advance to
ensure the clears are applied as expected
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21628>
Now that we have multiple sysval tables, implementing compute kernels --
including with indirect dispatch and load_num_workgroups -- is straightforward.
This patch adds the straightforward launch_grid implementation.
As usual needs UAPI support patches to actually do anything, but the relevant
compute tests are passing downstream.
It's not possible to properly test compute shaders support right now (pending
support for images), so we don't update the CAPs or features.txt here. This is
more about flushing out the piles of downstream patches we have (and getting
reviewed!) in preparation for cutting a downstream release soon.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21703>
The previous lowering was insufficient in two areas:
* No support for indirection. This is required for dynamically indexing into
UBOs, SSBOS, etc in OpenGL ES 3.2
* Only a single table supported. Multiple tables are required to implement
indirect dispatch/draws efficiently, in order to bind the indirect buffer as
uniforms.
The first problem is addressed here by reworking the lowering of
system values to happen in NIR, decoupled from the uniform register assignment
details, such that we can handle 1:n lowerings in a straightforward way.
Namely, indirect sysvals are lowered to indirect memory loads relative to the
base address of the sysval table, where the table address is itself pushed as a
(direct) sysval.
The second problem is addressed in this patch by generalizing to multiple
uniform tables.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21703>
Fragment shaders with side effects need to be lowered to ensure they execute for
all shaded pixels but no helper threads. Add a lowering pass to handle this.
Fixes dEQP-GLES31.functional.shaders.opaque_type_indexing.atomic_counter.const_literal_fragment
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21712>
This is required for address arithmetic to be lowered properly for compute
kernels, which may have u2u64 in the source NIR.
No shader-db changes (for GLES3.0).
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21430>
Most integer immediates are only 8-bit, but load/store instructions allow their
immediate offsets to be 16-bit instead. Take advantage of this in the optimizer.
This eliminates 36% of the instructions in
dEQP-GLES31.functional.ssbo.layout.random.all_shared_buffer.36, a fitting
percentage.
Insignificant effect on dEQP-GLES31.functional.ssbo.* performance... Only a
small % of our compile-time pie is actually spent in the backend anyway (as
opposed to NIR passes or GLSL IR).
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21430>
This avoids creating silly preambles that don't actually do anything except push
a constant that we could've inlined for cheaper anyway, since nir_opt_preamble's
cost model is sensitive to e.g. constant folding.
This avoids a pointless preamble in split-hell.
As a nice bonus, this also improves compile-time on address-heavy shaders. With
a release build, CPU time in dEQP-GLES31.functional.ssbo.* reduces from 12.87s
to 10.77... a 16% improvement is nothing to sneeze at.
shader-db results are mostly noise.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21430>
Useful both for ruling out issues with shader preambles as well as (in some
cases) making for a nicer reading experience of the compiled assembly.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21430>
It doesn't make sense, they're basically little compute kernel environments.
Noticed while debugging dEQP-GLES31.functional.fbo.no_attachments.multisample.*
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21710>
Kinda silly but fixes
dEQP-GLES31.functional.state_query.integer.max_framebuffer_samples_* which
queries the number of samples of a NONE format, required for
ARB_framebuffer_no_attachments to make sense.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21710>
Sample index and layer index are both 16-bits, even though they are zero
extended for compiler simplicity in some cases. In particular this means that 2D
MSAA arrays consume 6 half-regs for their coordinates, not 8. This is what the
IR translation (actually agx_nir_lower_texture) produces, we just need to fix
the calculation in agx_read_registers to agree.
Fixes validation failure in tests like
dEQP-GLES31.functional.texture.multisample.samples_4.use_texture_color_2d_array
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21708>
This gives our shifts SM5 behaviour at the cost of a little extra ALU. That way,
we match NIR's shifts.
This fixes unsoundness of GLSL expressions like "a << (b & 31)", where the &
would mistakenly get optimized away.
Closes: #8181
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reported-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21673>
vulkancts test dEQP-VK.wsi.direct_drm.surface.create_simulate_oom is failing
because in wsi_display_alloc_connector() function memory allocation for
connector is not checked for return NULL. create_simulate_oom test simulates
out of memory, hence memory allocation fails for connector and later when
tried to dereference connector program will segfault.
This patch fixes the dEQP-VK.wsi.direct_drm.surface.create_simulate_oom test
segfault issue by checking if connector is NULL afer memory allocation.
Cc: mesa-stable
Reviewed-by: Eric Engestrom <eric@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21701>
We'll need this to add a7xx support, since some of the regs are
different btwn a6xx and a7xx and reg variants are not supported with
the legacy reg builders.
Signed-off-by: Rob Clark <robdclark@chromium.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21706>
This NIR pass lowers stores in fragment shaders to:
if (!gl_HelperInvocaton) {
store();
}
This implements the API requirement that helper invocations do not have visible
side effects, and the lowering is required on any hardware that cannot directly
mask helper invocation's side effects. The pass was originally written for
Midgard (which has this issue) but is also needed for Asahi. Let's share the
code, and fix it while we're at it.
Changes from the Midgard pass:
1. Add an option to only lower atomics.
AGX hardware can mask helper invocations for "plain" stores but not for
atomics. Accordingly, the AGX compiler wants this lowering for atomics but
not store_global. By contrast, Midgard cannot mask any stores and needs the
lowering for all store intrinsics. Add an option to the common pass to
accommodate both cases.
This is an optimization for AGX. It is not required for correctness, this
lowering is always legal.
2. Fix dominance issues.
It's invalid to have NIR like
if ... {
ssa_1 = ...
}
foo ssa_1
Instead we need to rewrite as
if ... {
ssa_1 = ...
} else {
ssa_2 = undef
}
ssa_3 = phi ssa_1, ssa_2
foo ssa_3
By default, neither nir_validate nor the backends check this, so this doesn't
currently fix a (known) real bug. But it's still invalid and fails validation
with NIR_DEBUG=validate_ssa_dominance.
Fix this in lower_helper_writes for intrinsics that return data (atomics).
3. Assert that the pass is run only for fragment shaders. This encourages
backends to be judicious about which passes they call instead of just
throwing everything in a giant lower everything spaghetti.
Signed-off-by: Alyssa Rosenzweig <alyssa@rosenzweig.io>
Reviewed-by: Italo Nicola <italonicola@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21413>
These just don't seem to work. macOS falls back to eMRT here...
dEQP-GLES3.functional.draw_buffers_indexed.random.max_implementation_draw_buffers.13
from Fail -> Crash. Proper solution will come when we implement eMRT later on.
Signed-off-by: Asahi Lina <lina@asahilina.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21705>
Because we can have 2D views of 3D images, we want to consider the
slice info only with 3D *image views*.
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Fixes: 66e3ccbcac ("vulkan/runtime: store parameters of VK_EXT_sliced_view_of_3d")
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21376>
venus utilizes the host side shader cache.
This is a WA to generate shader cache files containing headers with
a unique cache id that will change based on host driver identifiers.
This allows fossilize replay to detect if the host side shader cache
is no longer up to date.
The shader cache is destroyed after creating the necessary files and
not utilized by venus.
Signed-off-by: Juston Li <justonli@google.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/21664>
