enums2names.py is only uses in one place. I propose to remove the -I
argument that is not strictly necessary as we can already get the
header name from the `-H` argument.
That modification is motivated by the need to help ninja-to-soong to
generate proper rule for the Android build system.
ninja-to-soong can't differenciate output file location and a string
matching the output file name.
Ref #14072
Acked-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37785>
Apparently various tessellation parameters come specified from
TESS_EVAL stage in GLSL while they come from the TESS_CTRL stage in
HLSL.
We switch to store the tesselation params more like shader_info with 0
values for unspecified fields. That let's us merge it with a simple OR
with values from from tcs/tes and the resulting merge can be used for
state programming.
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Fixes: a91e0e0d61 ("brw: add support for separate tessellation shader compilation")
Fixes: 50fd669294 ("anv: prep work for separate tessellation shaders")
Reviewed-by: Ivan Briano <ivan.briano@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37979>
It must not be larger than maxInlineUniformBlockSize.
Fixes VKCTS 1.4.4.0's
dEQP-VK.api.maintenance3_check.support_count_inline_uniform_block*.
Cc: mesa-stable
Signed-off-by: Valentine Burley <valentine.burley@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/38002>
Based on RADV.
The Vulkan spec says:
"If bindingCount is zero or if this structure is not included in
the pNext chain, the VkDescriptorBindingFlags for each descriptor
set layout binding is considered to be zero. Otherwise, the
descriptor set layout binding at
VkDescriptorSetLayoutCreateInfo::pBindings[i] uses the flags in
pBindingFlags[i]."
Fixes dEQP-VK.api.maintenance3_check.* in VKCTS 1.4.4.0.
Cc: mesa-stable
Signed-off-by: Valentine Burley <valentine.burley@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/38002>
Introduces an opt pass that attempts to optimize
load_barycentric_at_{sample,offset} with simpler load_barycentric_*
equivalents where possible, and optionally lowers
load_barycentric_at_sample to load_barycentric_at_offset with a position
derived from the sample ID instead.
Signed-off-by: Simon Perretta <simon.perretta@imgtec.com>
Reviewed-by: Marek Olšák <marek.olsak@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37658>
The subgroup lowering may generate new fp64 vector operations, so
ensure that those are lowered before calling nir_lower_doubles().
Issue spotted by Georg Lehmann.
Reviewed-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/38003>
hk_heap is called during command buffer recording, which may be
concurrent, so writing dev->heap without synchronization is a data race.
Signed-off-by: Olivia Lee <olivia.lee@collabora.com>
Fixes: 5bc8284816 ("hk: add Vulkan driver for Apple GPUs")
Reviewed-by: Mary Guillemard <mary.guillemard@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37973>
lowering bitsize before lowering idiv is silly, since then it forces us
down the software int32 division path instead of the much faster
int8/int16 lowered path. Relevant CTS tests:
dEQP-VK.spirv_assembly.type.scalar.i16.div_comp,
dEQP-VK.spirv_assembly.type.scalar.i8.rem_comp,
Go from:
SIMD8 shader: 46 instructions. 1 loops. 4716 cycles. 0:0 spills:fills
SIMD8 shader: 1008 instructions. 0 loops. 3600 cycles. 0:0 spills:fills, 8 sends
to:
SIMD8 shader: 17 instructions. 1 loops. 2556 cycles. 0:0 spills:fills
SIMD8 shader: 464 instructions. 0 loops. 1394 cycles. 0:0 spills:fills, 8 sends
No stats change on fossil-db (which has very little int8/int16 and even
less integer division, apparently).
Signed-off-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Reviewed-by: Ian Romanick <ian.d.romanick@intel.com>
Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37966>
Skip waiting/signaling on semaphores with stages not related
to a given subqueue
Reviewed-by: Lars-Ivar Hesselberg Simonsen <lars-ivar.simonsen@arm.com>
Reviewed-by: Yiwei Zhang <zzyiwei@chromium.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37810>
- rename vk_stage_to_subqueue_mask -> vk_stages_to_subqueue_mask
- handle stage masks instead of single stages.
- Add which sync scope it is reading to better reason with the mask semantics.
- Handle ALL_COMMANDS as well as TOP/BOTTOM (using sync scopes)
- add timestamp utility vk_stage_to_timestamp_subqueue_mask
Reviewed-by: Lars-Ivar Hesselberg Simonsen <lars-ivar.simonsen@arm.com>
Reviewed-by: Yiwei Zhang <zzyiwei@chromium.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37810>
Until now, the driver has been using a single set for internal state
while leaving maxBoundDescriptorSets to the remaining 15.
This gives us no room for optimizations of driver sets, which might
become an issue in the future.
To remedy this, we therefore reduce maxBoundDescriptorSets to 7. This
aligns with the proprietary driver and gives us the space to optimize
the driver sets.
We might increase this in the future if we see that we don't need all
the driver sets we now reserve.
Reviewed-by: John Anthony <john.anthony@arm.com>
Reviewed-by: Yiwei Zhang <zzyiwei@chromium.org>
Reviewed-by: Olivia Lee <olivia.lee@collabora.com>
Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Acked-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37978>
nir_lower_compute_system_values will attempt to lower
load_workgroup_size unless workgroup_size_variable is set. For precomp
shaders, the workgroup size is set statically for each entrypoint by
nir_precompiled_build_variant. Because we call
lower_compute_system_values early, it sets the workgroup size to zero.
Temporarily setting workgroup_size_variable while we are still
processing all the entrypoints together inhibits this.
Signed-off-by: Olivia Lee <olivia.lee@collabora.com>
Fixes: 20970bcd96 ("panfrost: Add base of OpenCL C infrastructure")
Reviewed-by: Eric R. Smith <eric.smith@collabora.com>
Reviewed-by: Mary Guillemard <mary.guillemard@collabora.com>
Acked-by: Alyssa Rosenzweig <alyssa.rosenzweig@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37799>
Turing and newer Nvidia cards can work with up to 8 discard rectangles
Reviewed-by: Mel Henning <mhenning@darkrefraction.com>
Signed-off-by: Lorenzo Rossi <git@rossilorenzo.dev>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/33476>
This can happen with (for example) 32x2 loads with
align_mul=4,align_offset=2.
This patch does bit_size=min(bit_size,bytes) to prevent num_components
from being 0.
Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Fixes: 52cd5f7e69 ("ac/nir_lower_mem_access_bit_sizes: Split unsupported shared memory instructions")
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37953>
Not sure about creating u64vec16 loads, but creating unaligned loads is
possible with opt_if_rewrite_uniform_uses.
Signed-off-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Georg Lehmann <dadschoorse@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37953>
The actual chances of this happening seem dubious, but the cleaned up
code seems nice. printf returns a value >= 0 on success, which is the
number of characters it writes a return < 0 means that an error
occurred, and then errno is set. Which negative value doesn't seem to be
specified, but it also seems unlikely that any implementation would
return `-MAX_INT`...
Anyway, this is fixed by converting the generic `print_repeated` to a
`print_separator` that avoids the need to do arithmetic at all by just
stopping the loop at 1 instead of 0, and then printing a newline.
CID: 1666497
CID: 1666256
CID: 1666531
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37746>
Coverity is pointing out that we should check this, and in reality if
this isn't what we expect the rest of the test is probably invalid
anyway.
CID: 1666504
CID: 1666544
CID: 1666552
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37750>
This was missed in the original maintenance9 MR.
Fixes the flakes in test
dEQP-VK.synchronization2.op.single_queue.event.write_ssbo_compute_read_ssbo_compute.buffer_16384_maintenance9
Fixes: 7692d3c0 ("nvk: Advertise VK_KHR_maintenance9")
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37964>
With untyped pointer we might get a deref_cast with a 0 ptr_stride. But we
were supposed to ignore the stride information on the pointer anyway, so
let's do that properly now.
Reviewed-by: Faith Ekstrand <faith.ekstrand@collabora.com>
Fixes: 05dca16143 ("nak: extract nir_intrinsic_cmat_load lowering into a function")
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/37941>
The normal encode pass writes batches to a section in build scratch
memory. Those batches contain information about the internal node and
the primitive nodes. The encoder is split to avoid the register
pressure of the compressor and maximize occupancy.
The compressor works in two passes because one pass can not guarantee
that every primitive node (except) has at least two triangles. This
guarantee is used to advertise a smaller acceleration structure size to
the application.
During compression, every invocation processes at most two triangles.
Groups of 8 invocations are used to support the maximum triangle count
of 16 that the hardware supports.
The first step of compression is loading the triangle(s). Shared
vertices are deduplicated early to avoid doing it in the compression
loop. The compression loop tries to add triangles to a list of triangles
until the computed node size needed for storing the triangles reaches
the hardware node size. For this, each invocation first deduplicates
vertices with the triangles that have already been picked. It then
computes the node size of the picked triangles plus the candidate
triangles of the current invocation. The invocation that computed the
smallest size is added to the list.
Because it may not be possible to fit every triangle into the same node,
there can be multiple hardware nodes which are written in parallel for
optimal performance. If there are no nodes with only one triangle, all
nodes are written. If there is, compression of the batch is aborted and
the index of the batch is written to build scratch memory. The second
compression pass will repeat the steps above but only for those aborted
batches. The nodes with only one triangle can and are now merged.
It can not be determined during box node encode which triangles will be
compressed together so the encoder also has to fix up the parent box
node's child infos.
Reviewed-by: Natalie Vock <natalie.vock@gmx.de>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/36965>
