The intrinsic produces a vec2, so let's honor that and avoid the weird
lowering to scalar and later reconstruction to vec2 when we find
load vulkan descriptor intrinsics.
It fixes tests like this (which require that we expose KHR_spirv_1_4):
dEQP-VK.spirv_assembly.instruction.spirv1p4.opptrequal.null_comparisons_ssbo_equal
that otherwise produce bad code that tries to access a vec2 from the result of
that intrinsic, leading to NIR validation errors.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11257>
We only support the VariablePointersStorageBuffer feature for now,
which is the only one that is mandatory, and for which we seem to
be passing all the relevant tests already.
Exposing the optional VariablePointers feature would require that
we support non-constant indexing on UBO/SSBO first.
Relevant CTS tests:
dEQP-VK.*pointer*
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11257>
I've seen this timing out on Zink, v3d, radeonsi, and Panfrost
periodically, but more and more lately. Try to minimise the damage by
just universally skipping it until someone comes up with a better idea.
Signed-off-by: Daniel Stone <daniels@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11267>
Report flakes in #videocore-ci channel.
v2 (Emma):
- Add glx@glx_arb_sync_control@timing.* as flakes.
Reviewed-by: Emma Anholt <emma@anholt.net>
Signed-off-by: Juan A. Suarez Romero <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11220>
As most of the development is done in arm64, execute the v3dv related
tests in this environment.
Reviewed-by: Emma Anholt <emma@anholt.net>
Signed-off-by: Juan A. Suarez Romero <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11220>
Make some of the jobs for vc4 and v3d to run automatically, in order to
identify more regressions.
Reviewed-by: Emma Anholt <emma@anholt.net>
Signed-off-by: Juan A. Suarez Romero <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11220>
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>
Make helper functions for all descriptor types and have them handle
all of the descriptor update so we can reuse them later to implement
template updates.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11213>
This extension is basically only wrapping SPV_KHR_storage_buffer_storage_class
which is entirely implemented in the SPIR-V frontend.
Relevant CTS tests:
dEQP-VK.glsl.opaque_type_indexing.ssbo_storage_buffer_decoration.*
dEQP-VK.spirv_assembly.*
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11184>
Fix defect reported by Coverity Scan.
Side effect in assertion (ASSERT_SIDE_EFFECT)
assignment_where_comparison_intended: Assignment deviceMask = 1U
has a side effect. This code will work differently in a non-debug
build.
Fixes: 234e1b7356 ("v3dv: implement VK_KHR_device_group")
Signed-off-by: Vinson Lee <vlee@freedesktop.org>
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11197>
We initially just copied on v3dv, just in case we needed to modify
it. One year later the code is exactly the same, so let's move it to
common.
This fix an additional issue, as we were not using NEON when building
v3d_tiling.c for v3dv.
v2:
* Add "#include util/u_box.h" at v3d_tiling.h, so we can't avoid
the need to include it on other places. (Juan and Iago)
Reviewed-by: Iago Toral Quiroga <itoral@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11121>
It seems our compiler already meets the requirements and we pass
all the relevant tests for this as far as I can see.
Relevant CTS tests:
dEQP-VK.ssbo.*relaxed*
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11159>
Some CTS tests use more than what we expose and other drivers
also seem to be exposing many more than us (in the order of thousands).
I don't think we want to expose a very large number since we use this
limit to size some arrays in the driver, but bumping it a bit over the
minimum of 4 required by the spec might be reasonable.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11159>
This also fixes the allocator used in v3dv_DestroyDevice.
v2: fix two more occurences of default_alloc (Roman Stratiienko)
Signed-off-by: Chia-I Wu <olvaffe@gmail.com>
Acked-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-by: Iago Toral Quiroga <itoral@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11117>
Always chain wsi_image_create_info to VkImageCreateInfo, which indicates
that the image is a wsi image and can be transitioned to/from
VK_IMAGE_LAYOUT_PRESENT_SRC_KHR.
Add prime_blit_buffer to the struct as well. When set, it indicates the
prime blit destination and implies that the image is a prime blit
source.
Signed-off-by: Chia-I Wu <olvaffe@gmail.com>
Reviewed-by: Yiwei Zhang <zzyiwei@chromium.org>
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10789>
This provides most of the implementation, but there are some
things we cannot enable until we improve of kernel submit
interface, namely:
We don't expose capacity to export SYNC_FD, although we do
have the implementation in place. This requires that we
improve our kernel interface and event wait implementation
first so we can cover the corner case where the application
submits a command buffer that includes a VkCmdWaitForEvents
and tries to export a SYNC_FD from its signal semaphores or
fence before it the event is signaled and the command buffer
is sent to the kernel for execution in full.
Likewise, we can't currently import semaphores. This is because
our current kernel submit interface can only take one syncobj.
We have been working around this so far by waiting on the last
syncobj produced from the device whenever we had to wait on any
semaphores (which is obviously suboptimal already), but this
won't work as soon as we allow importing external semaphores,
as those could (and would typically) be produced from a
different device.
Once we address the kernel bits, we should come back and enable
SYNC_FD exports as well as semaphore imports.
Relevant CTS tests:
dEQP-VK.api.external.fence.*
dEQP-VK.api.external.semaphore.*
dEQP-VK.synchronization.cross_instance.*
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11105>
We can (and should) close the descriptor immediately after the import.
Gets the following CTS test to pass without requiring to increase limits
for open file descriptors:
dEQP-VK.synchronization.basic.binary_semaphore.chain
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11105>
We were using the CT0CA (Control List Executor Current Address) and
CT0EA (Control List Executor End Address) registers, but that would
only wait for the CLE to reach the end of the list, but there could
still be things in the rest of the pipeline.
Even if that seems to work with the current simulator, the correct way
to do that is using the BFC (Binning Mode Flush Count) and RFC
(Rendering Mode Frame Count) registers instead.
In fact, this would be needed with a newer simulator snapshot, in
order to get the followint CTS tests working:
dEQP-VK.api.copy_and_blit.core.resolve_image.whole_array_image.4_bit
dEQP-VK.api.copy_and_blit.core.resolve_image.whole_array_image_one_region.4_bit
dEQP-VK.api.copy_and_blit.core.resolve_image.whole_copy_before_resolving.4_bit
dEQP-VK.api.device_init.create_instance_device_intentional_alloc_fail
dEQP-VK.api.image_clearing.core.clear_color_image.1d.optimal.multiple_layers.r32g32_uint
dEQP-VK.api.image_clearing.core.clear_color_image.1d.optimal.remaining_array_layers_twostep.r16_sint
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11039>
Until now we were waiting until having a dispatch current and/or
queued. But that would only wait for all shaders to have started, it
won't wait for them to have finished.
With this commit we wait until the NUM_COMPLETED_JOBS (that in spite
of that name, it is about dispatches) field got increased.
This is in general safest, and it is needed after the latest simulator
update to get CTS tests like the following ones working:
dEQP-VK.compute.basic.copy_ssbo_multiple_invocations
dEQP-VK.compute.basic.copy_ssbo_single_invocation
dEQP-VK.compute.basic.ssbo_rw_single_invocation
dEQP-VK.compute.basic.ssbo_unsized_arr_single_invocation
dEQP-VK.compute.basic.ubo_to_ssbo_multiple_invocations
dEQP-VK.compute.basic.ubo_to_ssbo_single_invocation
v2 (from Juan feedback):
* Clarify JOBS vs DISPATCHES
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11039>
Current code just assumes that flushes are instant, as simulator
doesn't really model the caches. So right now we have just an assert
that the flush has been done.
But that can change on the future, so let's change the assert for a
wait.
Note that for the l1t case we are writing on the field TMUWCF. So I
understand that then we need to wait for TMUWCF_SET, even if the
previous code was using L2TFLS_SET.
This also happpens on the kernel side. We need to check if this was a
typo on the kernel side.
v2 (from Juan feedback)
* Add comment about the TMUWCF vs L2TFLS difference between this
commit and the kernel.
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11039>
So far we were not capturing any HUB interrupt, just core. This could
be a problem if any is fired, as we could enter on an infinite
loop. With this commit we start to capture them. So we split v3d_isr
into core and hub interrupt handling.
As reference we capture the same HUB interrupts that we capture on the
v3d kernel support.
It is worth to note that all those are mostly untested. Now with both
opengl/vulkan driver being stable we were not able to raise those
interrupts.
v2 (Juan feedback):
* Just one V3D_VERSION >= 41 block, more readable
* Assert that the core is 0 at v3d_isr_core (we don't handle
multi-core right now).
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11039>
We were using the num_components to infer it, but in the end it is
VEC2 for CMPXCHG and 32BIT for anything else.
This doesn't affect any test with the real hw, but fixes an assert
with the last version of the simulator.
Reviewed-by: Juan A. Suarez <jasuarez@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11039>
There are some interactions between these two extensions that need to be
implemented when both are supported. Particularly:
1. Applications can create images that will be bound to swapchain memory
by passing a VkImageSwapchainCreateInfoKHR in the pNext chain
of VkImageCreateInfo. In this case we need to make sure that the
created image takes some of its parameters from the underlying
swapchain.
2. Applications can bind memory from a swapchain image to a VkImage
by passing a VkBindImageMemorySwapchainInfoKHR in the pNext chain
of VkBindImageMemoryInfo.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11037>
This was added with VK_KHR_device_group and allows users to specify
a base offset that will be automatically added to gl_WorkGroupID.
Unfortunately, V3D doesn't support this natively, so we need to add
the base to the workgroup id generated by hardware manually. For this,
we inject add instructions that source from a QUNIFORM that will
retrieve the actual dispatch base from the compute job when it is
dispatched.
Since a compute shader can be dispatched with CmdDispatch and/or
CmdDispatchBase, we always need to add these additional add
instructions and use a base of (0,0,0) for regular dispatches.
Since we don't support any version of OpenGL with this dispatch
base functionality we can avoid the extra instructions there.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11037>
This one is not implemented in the common dispatch handler in terms
of its KHR_get_memory_requirements2 version, so the driver needs
to implement it.
Fixes: d87afc1acc ('v3dv: implement VK_KHR_get_memory_requirements2')
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11038>
We don't have any special restrictions associated with the number
of descriptors in a set other than maybe not exceeding what we can
put in a single memory allocation, so in practice, applications will
be limited by the per-stage contraints defined by other Vulkan limits.
Reviewed-by: Alejandro Piñeiro <apinheiro@igalia.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10970>
