KHR-GL32.packed_pixels.pbo_rectangle.r16i on zink on lavapipe
ends up using a pbo that does an SINT image write. This was producing
truncated rather than clamped values.
Fix the calculations for 8/16-bit signed ints to clamp not truncate.
Fixes: 13e5f331db ("gallivm/nir: fix image store conversions")
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13187>
This uses pmulhrsw avx2 and ssse3 variants. It fixes the
precision of texture filtering calculations.
However it does leave these paths inaccurate on platforms
that don't support it.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Reviewed-by: Adam Jackson <ajax@redhat.com>
Acked-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13075>
Again if you get passed an invoc but the exec mask has the
active lane somewhere other than at 0, then if we have an
invoc we should find the active lane and extract the value
from invoc rather than using the idx.
This fixes a bunch of VK 1.2 subgroup tests once 1.2 is enabled:
dEQP-VK.subgroups.ballot_broadcast.compute.subgroupbroadcast_nonconst*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Acked-by: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12953>
The fails will be addressed later.
This adds a fail in GLSL compiler that is due to a workaround
that fails when fp16 constants are lowered
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11816>
Otherwise the inf translations don't seem to work, and the VK CTS
fails
Fixes VK CTS dEQP-VK.spirv_assembly.instruction.graphics.float16.arithmetic*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11816>
Some VK CTS tests are topping this out around 76, increase it to 80 for now.
Fixes:
dEQP-VK.spirv_assembly.instruction.graphics.float16.opvectorshuffle.*44*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11816>
This is an initial patch that is needed for OpenCL and Vulkan
support for proper 16-bit floats.
This doesn't enable the cap bit yet
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11816>
Get comfy.
llvmpipe coroutines have a stack frame. This is created by hooking
in malloc and coro.alloc and coro.size intrinsics.
LLVM has an CoroElide pass that is meant to allow that stack frame
to be done as an alloca in the caller instead of using the malloc path.
The CoroElide pass relies on the coroutine being inlined (fixed that).
The CoroElide pass relies on there being a direct connect between
coro.destroy(i8 *arg) and arg = coro.begin(id). However due to the
way the compute shaders are launched, there is no way to ensure that
link. Fixing the CoroElide pass seems quite difficult, I considered
having a force CoroElide always flag to make it dtrt, however I'm not
sure how ugly that would end up.
My first attempt tried to preallocate the stacks at a fixed size,
this turned out to be naive as the stack frame size was not sized
like I expected. Instead the first coro to run allocs enough for
everyone, so avoid the massive amounts of small allocations.
This remove coro malloc from a lot of profiles and shaves another 30s
or so from OpenCL ./conversions/test_conversions uchar_uin
(from 4.40m to just under 4m on my ryzen 7 1800x)
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12432>
Although the index has to be dynamically uniform, if we don't ever
execute a few lanes then we'll have 0, so it important to read the
ssbo index from the first active lane.
Just loop over them all.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12689>
16bit integer support also implies using 16-bit results when sampling
textures.
Because we're returning the results in float SSA values instead of int,
we need to bitcast back to integers before truncating the values.
Fixes: 00ff60f799 ("gallivm: add 16-bit integer support")
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12413>
This fixes integer division for non-32bit but also fixes the
32-bit case where INT_MIN/-1 causes an exception.
Fixes CL CTS
./integer_ops/test_integer_ops quick_long_math
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12285>
The frontend can't handle variable block sizes properly,
so just lower it here in the backend.
Fixes CTS basic local_linear_id + get_linear_ids
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12285>
These two bits were added in 2012, but never got wired up. Let's cut our
losses, and remove them again. 9 years unused seems sufficient.
While we're at it, remove reduction_mode from the hacks-section, because
this isn't a hack at all, rather normal state.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12240>
There's no need for us to do this lowering ourselves while emitting
code, when there's already a helper that can do this for us that we're
even using. Let's just set the right flag, and not worry about
projectors any more.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Reviewed-by: Emma Anholt <emma@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12126>
The VK spec requires that derivatives are mapped to the face, and given
that the face is per pixel that implies that we need per-pixel
derivatives. This also seems to match hardware behavior.
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10219>
It's non-conformant for GL and Vulkan, and not expected to be a desired
quality/speed tradeoff for applications using llvmpipe. The option is
left in place using GALLIVM_PERF=brilinear.
Causes minor rasterization changes in our traces:
- improved sharpness of the trees in CS
- more consistent mipmap filtering of the ground in CS:Source, STK and the
RaytracedShadows demo.
- changing some aliasing on shadow maps in 0ad (not consistently
visibly better or worse)
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10219>
GALLIVM_PERF_NO_OPT is "belond" to GALLIVM_PERF environment variable,
not GALLIVM_DEBUG. when GALLIVM_PERF=nopt is passed to llvmpipe,
"-sroa -early-cse -simplifycfg -reassociate -constprop -instcombine -gvn"
should not be printed. Those llvm optimation techniques should only be
printed when the default(-O2) optimation is enabled.
$ GALLIVM_PERF=nopt GALLIVM_DEBUG=dumpbc glmark2
Before apply this patch:
Invoke as "opt -sroa -early-cse -simplifycfg -reassociate -mem2reg -constprop -instcombine -gvn ir_setup_variant_0.bc | llc -O2 [-mcpu=<-mcpu option>] [-mattr=<-mattr option(s)>]"
After apply this patch:
Invoke as "opt -mem2reg ir_fs304_variant0.bc | llc -O0 [-mcpu=<-mcpu option>] [-mattr=<-mattr option(s)>]"
Reviewed-by: Adam Jackson <ajax@redhat.com>
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Signed-off-by: suijingfeng <suijingfeng@loongson.cn>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11994>
When the dot product uses a source which can be optimized to a scalar,
after a bunch of nir optimization steps the source to fsum will be a scalar
with a x replicate swizzle. Hence nir_src_num_components is just 1 and the
fsum was just a no-op which is not correct. Arguably this could be optimized
a bit better, but just determine the number of addends by using nir_op_infos
instead (the operand fetch was fixed already by 39a938ecf4 doing the same).
Fixes: 4eb0475b5a ("gallivm/nir: add fsum support")
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Reviewed-by: Brian Paul <brianp@vmware.com>
Reviewed-by: Dave Airlie <airlied@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12066>
Only fragment and some compute shaders support implicit derivatives.
They're totally meaningless without helper invocations and some
understanding of the dispatch pattern. We've got code to lower
nir_texop_tex in these shader stages to use an explicit derivative of 0
but it was pretty badly broken:
1. It only handled nir_texop_tex, not nir_texop_txb or nir_texop_lod.
2. It didn't take min_lod into account
3. It was conflated with adding a missing LOD parameter to opcodes
which expect one such as nir_texop_txf. While not really a bug,
this does make it way harder to reason about the code.
4. Unless you set a flag (which most drivers don't), it left the
opcode nir_texop_tex instead of nir_texop_txl which it should have
been.
This reworks it to go through roughly the same path as other LOD
lowering only with a constant lod of 0 instead of calling out to
nir_texop_lod. We also get rid of the lower_tex_without_implicit_lod
flag because most drivers set it and those that don't are probably
subtly broken. If someone really wants to get nir_texop_tex in their
vertex shaders, they can write a new patch to add the flag back in.
Fixes: e382890e25 "nir: set default lod to texture opcodes that..."
Fixes: d5ac5d6e83 "nir: Add option to lower tex to txl when..."
Reviewed-by: Connor Abbott <cwabbott0@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11775>
VK CTS has a test that does -256 for sample index but since we use
signed >= we fail and write to random memory, changing to uint here
for comparisons for out of bounds and address calcs should ensure,
we stay within bounds even for negative vals.
Remove unused int_type/bld while here.
Fixes dEQP-VK.texture.multisample.invalid_sample_index.sample_count_4
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11913>
Use the newly added wrapper lp_set_module_stack_alignment_override() to
achieve the same end result.
Signed-off-by: Kai Wasserbäch <kai@dev.carbon-project.org>
Closes: mesa/mesa#4906
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11940>
We need it in the next commit to replace setting the stack alignment on
i386 with LLVM >= 13 through the TargetOption::StackAlignmentOverride,
which was removed in the upstream commit
<3787ee4571>.
Unfortunately Module::setOverrideStackAlignment() is not available
through the C API and we need to wrap it ourselves.
Signed-off-by: Kai Wasserbäch <kai@dev.carbon-project.org>
Reference: mesa/mesa#4906
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11940>
This change adds:
- an alternative rasterizer, which rasterizes bins in a left->right &
top->bottom linear fashion;
- triangle -> rectangle detection;
- 1:1 blit detection;
- a special TGSI -> LLVM IR code generation that uses 8-bit SSE integers
in AoS fashion (as opposed to 32bits floats.)
Altogether these changes yield a 2x to 3x performance improvement for 2D
workloads. It was designed to render Windows 7 Aero and other Windows
built-in 3D applications (like Windows Media Player, Internet Explorer
11, UWP applications) with minimum CPU utilization, but it should be
generally applicable to other 2D-on-3D applications, like desktop
compositors, HTML browsers, 3D based UI toolkits, etc.
This was mostly the brainchild of Keith Whitwell back in 2010. I wrote
TGSI -> AoS translation. And many others added bug-fixes and
enhancements over the years: Roland Scheidegger, Brian Paul, and James
Benton.
Known issues:
- piglit spec@!opengl 1.1@quad-invariance will warn that "left and right
half should match" due to rounding error difference
- These optimized paths to kick in is that depth-buffer must not be
used, so some applications which want to benefit from these improvements
might need to be modified to ensure they use painter's algorithm instead
of depth-buffers.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Reviewed-by: Brian Paul <brianp@vmware.com>
Acked-by: Keith Whitwell <keithw@vmware.com>
v2: Incorporate Dave Airlie feedback: cleanup LP_DEBUG_xx; shrink 3+
empty lines.
v3: silence unused var warning, adapt to new upstream code (point setup)
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11969>
ls3a4000 and ls2k1000 cpu is mips64r5 compatible with MSA SIMD
instruction set implemented, while ls3a3000 is mips64r2 compatible only.
Due to lacking llvm support for loongson CPU, llvm::sys::getHostCPUName().
return "generic" on all loongson mips CPU.
So we override the MCPU to mips64r5 if MSA is implemented, feedback to
mips64r2 for all other ordinaries.
Reviewed-by: Adam Jackson <ajax@redhat.com>
Signed-off-by: suijingfeng <suijingfeng@loongson.cn>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11955>
This is a port of the softpipe anisotropic filtering
to llvmpipe. It should produce pretty similiar results.
This contains the proposed fix to the softpipe calculating
dq after scaling.
It also contains a number of other fixes around vector lengths
etc caught during test.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8804>
The anisotropic filtering needs access to a table of weights,
to make the calculations easier. This routes the table
through from shader parameter and makes it available for the
sampler code.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8804>
Qualcomm has a mode with a subgroup size of 128, so just emitting larger
integer operations and then lowering them later isn't an option. This
makes the pass able to handle the lowering itself, so that we don't have
to go down to 64-thread wavefronts when ballots are used.
(The GLSL and legacy SPIR-V extensions only support a maximum of 64
threads, but I guess we'll cross that bridge when we come to it...)
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6752>
Be consistent with other usages in Vulkan and SPIR-V, and the recently
added workgroup_size field.
Acked-by: Emma Anholt <emma@anholt.net>
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Acked-by: Timur Kristóf <timur.kristof@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11190>
We can do stuff like this:
vec1 32 ssa_207 = fsum3 ssa_209.xxx
In this case, we'd end up not swizzling in get_alu_src, and reading
components out-of-bounds, which LLVM isn't very happy about, and thus
takes punitive actions, in the form of a segfault.
We don't want that, and we already know from the opcode what the
component counts should be here.
Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10690>
We have to unwind the lod into the scalar path correctly.
Fixes a crash with renderdoc demo
Fixes: e168d148d7 ("gallivm/nir: handle non-uniform texture offsets")
Reviewed-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10676>
Since the second source is always a constant that is known to be a
number, this should have the same performance as
GALLIVM_NAN_BEHAVIOR_UNDEFINED.
A lofty goal is to eventually remove GALLIVM_NAN_BEHAVIOR_UNDEFINED.
There's still a lot of (mostly implicit) users, and I don't feel like
tackling that right now. :)
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10532>
If it is known that one of the source must be a number, then the (more
efficient) GALLIVM_NAN_RETURN_OTHER_SECOND_NONNAN path can be used.
v2: s/know to be/known to be/. Noticed by Roland.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10532>
Like softpipe in mesa!10419, llvmpipe suffers from improper handling
of NaN in nir_op_fmax and nir_op_fmin. nir_op_fsat is already handled
correctly. OpenCL strictly requires the "NaN cleansing" behavior, so
all of the functionality is in place. Just make the graphics APIs use
the OpenCL path.
The majority of the possible performance penalty incurred here should
be resolved in the next commit.
v2: Add updated checksum for bgfx/39-assao.rdc trace. Rendering goes
from mostly garbage to looking correct to me.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10532>
Without this, we're left with a bunch of medium-precision residue, like
i2imp. And to avoid that the lowering intrudces left-over instructions,
also run other lowering passes, similar to what other drivers do.
Reviewed-by: Dave Airlie <airlied@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10185>
One exception is src/amd/addrlib/, for which -Wimplicit-fallthrough is
explicitly disabled.
Reviewed-by: Eric Anholt <eric@anholt.net>
Reviewed-by: Alyssa Rosenzweig <alyssa@collabora.com>
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Reviewed-by: Gert Wollny <gert.wollny@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10220>
Need to bump up the size of texture target bitfield for MSVC.
Fixes: 0ce7c4a7c9 ("gallivm: Use the proper enum for the texture target bitfield.")
Reviewed-by: Neha Bhende <bhenden@vmware.com>
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9928>
8/16-bit storage requires ubo loads for the smaller types,
fix the ubo loading and bounds checking.
Acked-by: Adam Jackson <ajax@redhat.com>
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9643>
In release builds, there should be no change, but in debug builds the
assert will help us catch undefined behavior resulting from using
util_cpu_caps before it is initialized.
With fix for u_half_test for MSVC from Jesse Natalie squashed in.
Signed-off-by: Rob Clark <robdclark@chromium.org>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9266>
../src/gallium/auxiliary/gallivm/lp_bld_sample_soa.c: In function ‘lp_build_sample_image_linear’:
../src/gallium/auxiliary/gallivm/lp_bld_sample_soa.c:1412:24: warning: ‘have_corners’ may be used uninitialized in this function [-Wmaybe-uninitialized]
1412 | have_corner = LLVMBuildLoad(builder, have_corners, "");
I don't think you can get here and really not have it initialized, but
if you _could_ then it would clearly be wrong to do anything, so do
nothing instead.
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8724>
../src/gallium/auxiliary/gallivm/lp_bld_arit.c: In function ‘lp_build_round_arch’:
../src/gallium/auxiliary/gallivm/lp_bld_arit.c:2042:7: warning: ‘intrinsic_root’ may be used uninitialized in this function [-Wmaybe-uninitialized]
2042 | lp_format_intrinsic(intrinsic, sizeof intrinsic, intrinsic_root, bld->vec_type);
Can't happen, mark it unreachable.
Reviewed-by: Eric Anholt <eric@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8724>
The tgsi.rst was not very clear but didn't indicate that they were
supported, and llvmpipe only supported double abs and not negate.
glsl_to_tgsi never generated them, and neither did the new nir_to_tgsi, so
just garbage collect it.
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Gert Wollny <gert.wollny@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8369>
llvmpipe was reporting 32 max const buffers, while sizing its arrays to 16
according to gallivm's #define.
Fixes: 1d35f77228 ("gallivm,llvmpipe,draw: Support multiple constant buffers.")
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7695>
We rename it to "modes" to make it clear that it may contain more than
one mode and adjust all the uses of nir_deref_instr::modes to attempt to
handle multiple modes.
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Caio Marcelo de Oliveira Filho <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6332>
NIR derefs currently have exactly one variable mode. This is about to
change so we can handle OpenCL generic pointers. In order to transition
safely, we need to audit every deref->mode check. This commit adds a
set of helpers that provide more nuanced mode checks and converts most
of NIR to use them.
For simple cases, we add nir_deref_mode_is and nir_deref_mode_is_one_of
helpers. These can be used in passes which don't have to bother with
generic pointers and just want to know what mode a thing is. If the
pass ever encounters generic pointers in a way that this check would be
unsafe, it will assert-fail to alert developers that they need to think
harder about things and fix the pass.
For more complex passes which require a more nuanced understanding of
modes, we add nir_deref_mode_may_be and nir_deref_mode_must_be helpers
which accurately describe the compiler's best knowledge about the given
deref. Unfortunately, we may not be able to exactly identify the mode
in a generic pointers scenario so we have to be very careful when we use
these. Conversion of these passes is left to later commits.
For the case of mass lowering of a particular mode (nir_lower_explicit_io
is one good example), we add nir_deref_mode_is_in_set. This is also
pretty assert-happy like nir_deref_mode_is but is for a set containment
comparison on deref modes where you expect the deref to either be all-in
or all-out.
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Caio Marcelo de Oliveira Filho <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6332>
This code works but isn't entirely correct, for a dvec3 it would
fetch loc 0,1 2,3 4,5 but really each loc only has 4 entries,
instead catch this and read loc 0,1 2,3 loc+1 0,1
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7416>
Due to flow control we can end up with random values in here having
side effects.
This fixes a crash in gtk4-demo.
Fixes: 44a6b0107b ("gallivm: add nir->llvm translation (v2)")
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7327>
Scratch space is per-thread space, so allocate the scratch size
* vector width, and add a per-thread base offset to each
load/store.
This is needed for OpenCL private memory space
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7304>
This reverts commit 4fb2eddfdf.
This reverts commit 7a1deb16f8.
This reverts commit 2b6a172343.
This reverts commit 5af81393e4.
This reverts commit 87900afe5b.
A couple of problems were discovered after this series was merged that
cause breakage in different configurations:
(1) It seems that using -mf16c also enables AVX, leading to SIGILL on
platforms that do not support AVX.
(2) Since clang only warns about unknown flags, and as I understand
it Meson's handling in cc.has_argument() is broken, the F16C code is
wrongly enabled when clang is used, even for example on ARM, leading
to a compilation error.
Closes: https://gitlab.freedesktop.org/mesa/mesa/-/issues/3583
Reviewed-by: Eric Anholt <eric@anholt.net>
Reviewed-by: Rob Clark <robdclark@chromium.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6969>
Make the driver only read values for the active lanes,
otherwise it can cause unwanted oob accesses that aren't
the apps fault.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6954>
This makes it explicit that this intrinsic is only for SSBOs. For the
v3dv driver, we'll be adding a get_ubo_size intrinsic and we want to be
able to distinguish between the two.
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Iago Toral Quiroga <itoral@igalia.com>
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6812>
GLSL lowers packhalf2x16 itself, but for SPIRV we don't have that
option.
For packing when NIR lowers it uses f2f16
and for unpack it needs the casting and f2f32
Fixes:
dEQP-VK.glsl.builtin.function.pack_unpack.packhalf2x16*
dEQP-VK.glsl.builtin.function.pack_unpack.unpackhalf2x16*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6381>
This was suggested by Roland, and fixes stencil images.
Fixes:
dEQP-VK.renderpass.dedicated_allocation.formats.d24_unorm_s8_uint.*
dEQP-VK.renderpass.dedicated_allocation.formats.d32_sfloat_s8_uint.*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6381>
It was always fneu but naming it fne causes confusion from time to time. So
lets rename it. Later we also want to add other unordered and fne, this is
a smaller preparation for that.
Signed-off-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by: Bas Nieuwenhuizen <bas@basnieuwenhuizen.nl>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-by: Connor Abbott <cwabbott0@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6377>
vulkan has these separate, should be fine for GL as well as
the values will be the same anyways.
Fixes:
dEQP-VK.binding_model.shader_access.secondary_cmd_buf.uniform_texel_buffer.*
dEQP-VK.binding_model.descriptorset_random.sets4.noarray.ubolimitlow.sbolimitlow.sampledimglow.outimgtexlow.noiub.uab.vert.noia*
dEQP-VK.binding_model.descriptor_copy.graphics.uniform_texel_buffer.*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6339>
This fixes a lot of tessellation shaders, since tess factors
get emitted with spir-v now.
Fixes a bunch of:
dEQP-VK.tessellation*
along with 7000 others.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6339>
The lod query doesn't take a layer, but the code tries to use one,
detect lodq and don't use a layer in those cases.
There appears to be no GL tests for this behaviour, but the vulkan
CTS hits it.
Fixes:
dEQP-VK.glsl.texture_functions.query.texturequerylod.*
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6339>
This convets the push constant load to a ubo load against the 0 constant buffer,
which is where the vallium layer will bind things.
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6082>
A single format either had the float, the sint, or the uint version.
Making the dst be void * lets us store them in the same slot and not have
logic in the callers to call the right one.
-6kb on gallium drivers
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6305>
All drivers that support mediump lowering should support 16BIT_TEMPS,
but some do not also want 16b consts to be lowered. Replace the pipe
cap in preperation to remove LowerPrecisionTemporaries.
Note: also updates reference checksums for the arm64_a630_traces job,
due to lowering more to 16b
Signed-off-by: Rob Clark <robdclark@chromium.org>
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@collabora.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Reviewed-by: Eric Anholt <eric@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6189>
Dave Airlie
Dave Airlie
Pavel Asyutchenko
Erik Faye-Lund
Emma Anholt
suijingfeng
Roland Scheidegger
Jason Ekstrand
Kai Wasserbäch
Jose Fonseca
Connor Abbott
Michel Zou
Caio Marcelo de Oliveira Filho
Mike Blumenkrantz
Timur Kristóf
Adam Jackson
Jesse Natalie
Ian Romanick
Michel Dänzer
Marek Olšák
Charmaine Lee
Rob Clark
Eric Anholt
Pierre-Eric Pelloux-Prayer
Tapani Pälli
Matt Turner
Karol Herbst