Commit Graph

69 Commits

Author SHA1 Message Date
Karol Herbst 9c5fd100cc nir: add a nir_remove_non_entrypoints helper
This code just got duplicated a lot. There is still more, but the
remaining instances do a bit more than just removing other functions.

Signed-off-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by: Alyssa Rosenzweig <alyssa@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/16348>
2022-05-10 03:37:44 +00:00
Erik Faye-Lund f8fe225840 microsoft/clc: fixup indentation
Acked-by: Boris Brezillon <boris.brezillon@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/16079>
2022-04-29 07:08:57 +00:00
Jason Ekstrand 319d87846c nir,microsoft: Move scale_fdiv into a common NIR pass
While we're at it, convert to nir_shader_instructions_pass() to get rid
of some boilerplate and get metadata correct.

Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/15983>
2022-04-16 02:10:25 +00:00
Erik Faye-Lund 0998621496 clc/tests: use dxil_validator
Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/15751>
2022-04-07 00:00:45 +00:00
Jesse Natalie 7f9ac5bba2 d3d12: Support dynamic UBO/SSBO indexing
Reviewed-by: Sil Vilerino <sivileri@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14624>
2022-01-21 23:08:26 +00:00
Jesse Natalie fd50ef046b microsoft/compiler: Move workgroup_size lowering from clc
It doesn't depend on the clc data being provided externally, so no
need to tie it there, we can re-use it for GL and Vulkan compute.

Reviewed-by: Sil Vilerino <sivileri@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14367>
2022-01-11 01:36:56 +00:00
Jesse Natalie 8a4b443c5b microsoft/compiler: Change vulkan_environment bool to an enum
There are currently 3 different "environments" supported by this backend,
and they have slightly different semantics for how resources are accessed,
which is only going to get a little weirder when GL images start getting
supported. To try to make things less confusing, use an explicit enum
with heavy documentation on what's different between them.

Reviewed-by: Sil Vilerino <sivileri@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14342>
2022-01-07 03:31:16 +00:00
Jesse Natalie f0e5bc228c microsoft/clc: Add a test for arg metadata
Reviewed-by: Karol Herbst <kherbst@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13759>
2021-11-15 07:47:06 -08:00
Jesse Natalie 8d3a3e7a00 microsoft/compiler: Use textures for SRVs
After running the (renamed) dxil_nir_split_typed_samplers pass, the
shader will have either:
* Textures, which map to D3D SRVs
* Bare samplers, which map to D3D bare samplers
* Images, which map to D3D UAVs

There shouldn't be any remaining samplers with type information

Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13390>
2021-11-02 11:02:22 -07:00
Jason Ekstrand 956199e870 nir: s/nir_var_mem_image/nir_var_image/g
We typically use nir_var_mem_* for stuff that has an explicit byte-based
memory layout.  Images are opaque.

Reviewed-by: Caio Marcelo de Oliveira Filho <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13386>
2021-10-16 03:47:10 +00:00
Dylan Baker e73096bd6d meson: use gtest protocol for gtest based tests when possible
With the `gtest` protocol meson will add some extra arguments to the
test to generate better junit results, which may be useful. This
protocol is only available in meson 0.55.0+, so keep using the default
`exitcode` protocol for meson older than that.

Reviewed-by: Caio Marcelo de Oliveira Filho <caio.oliveira@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/8484>
2021-10-16 03:22:24 +00:00
Jesse Natalie 9601556079 microsoft/clc: Images use nir_var_mem_image
The only big change is that lower_vars_to_explicit no longer assigns
a driver_location for images. That means that the storage for the
format/order loads is no longer implicitly "allocated" in the middle
of the kernel args. Instead, manually add the storage for that to the end
of the input args buffer.

Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/4743>
2021-10-15 14:58:56 +00:00
Lionel Landwerlin 3c8c817ae7 clc: add allowed extension for compile parameter
The LLVM-SPIRV translator can include a bunch of capabilities into the
generated SPIRV which is not what you always want. That include
internal Intel specific capabilities from the translator.

v2: Rename options
    Fixup checks (Jesse)

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13113>
2021-10-03 19:32:54 +00:00
Jason Ekstrand 1506ea2ecb Move a bunch of the CLC stuff from src/microsoft to common code
The D3D12-specific stuff isn't useful to have in common code but all the
stuff to invoke clang really should be common.

v2: Rebase (Lionel)

v3: Define a new clc_libclc_new_dxil() entrypoint to create a clc
    context with DXIL nir_options (Jesse)

v4: Fixup meson build (Lionel)

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Dylan Baker <dylan@pnwbakers.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9156>
2021-09-30 07:09:08 +00:00
Lionel Landwerlin a9c49a0541 meson: extract libversion checks from clc & clover
The src/microsoft/clc/meson.build was assuming to be run only on
Windows. That's about to change.

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Dylan Baker <dylan@pnwbakers.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9156>
2021-09-30 07:09:08 +00:00
Lionel Landwerlin 2bc4650b45 microsoft/clc: fix compiler warning on uninitiailzed variable use
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Dylan Baker <dylan@pnwbakers.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9156>
2021-09-30 07:09:08 +00:00
Lionel Landwerlin 1d7a5196b1 microsoft/clc: drop MSVC specific function
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Dylan Baker <dylan@pnwbakers.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9156>
2021-09-30 07:09:08 +00:00
Lionel Landwerlin 32358b0335 microsoft/clc: drop LLVM dependency to version < 12
Prior to LLVM 12, SmallVector requires 2 template arguments.

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Fixes: c4755a7c32 ("microsoft/clc: Support SPIR intermediates in the compilation APIs")
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13083>
2021-09-28 13:29:56 +00:00
Lionel Landwerlin 1c92d355c6 clc: use the defined version for the parser
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13018>
2021-09-24 12:33:19 +00:00
Lionel Landwerlin dce8870f64 microsoft/clc: small compile fix on Linux
Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13018>
2021-09-24 12:33:19 +00:00
Jesse Natalie 7aa060ec38 microsoft/clc: Add a test for specializing via SPIRV-Tools
Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie 068c6b5a37 microsoft/clc: Add API to independently specialize SPIR-V
We need the ability to specialize unlinked SPIR-V, so use SPIR-V tools
to specialize prior to linking.

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie c50bbf1f28 microsoft/clc: Support passing specialization consts to spirv_to_nir
Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie a699844ffb microsoft/clc: Parse SPIR-V specialization consts into metadata
We need to be able to validate at the API that set specialization consts
have a valid ID and the value is the correct size.

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie c4755a7c32 microsoft/clc: Support SPIR intermediates in the compilation APIs
We can now export SPIR (mainly just for testing) or import SPIR and
convert to SPIR-V.

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie 91e08312d8 microsoft/clc: Split clc_object and rename entrypoints
clc_object was overloaded, containing SPIR-V binary and metadata,
and it would only sometimes contain metadata (after linking). Split
it into a more generic clc_binary class which holds some type of data
(the kind depends on where it came from), and clc_metadata which can
be independently parsed on compiled or linked data.

Rename a couple entrypoints to be more explicit about what they're
actually transforming (c_to_spirv, link_spirv, spirv_to_dxil).

Add a logger to SPIR-V binary parsing so it can report errors on app-
provided SPIR-V.

Re-order helper function parameters to be more consistent (out params last).

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie c3666eec7e microsoft/clc: Stop heap-allocating tiny fixed-size transparent structs
The caller can allocate these however they want. They don't need
independent allocations. Removes some unnecessary failure handling.

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie 27467700e9 microsoft/clc: Clean up clc_context
1. Rename it to libclc to match what it is
2. Make it opaque externally
3. Remove it from non-essential entrypoints (compile/link)

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Jesse Natalie 24b096035f microsoft/clc: Rename compiler DLL to clon12compiler
I no longer think there's value in trying to share code on disk here between
CL and GL at runtime, especially since we're now just building GL without LLVM

Acked-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10322>
2021-09-23 13:06:41 +00:00
Marcin Ślusarz 17a61ec541 microsoft/clc: use nir_shader_instructions_pass in clc_nir_dedupe_const_samplers
Changes:
- nir_metadata_preserve(..., nir_metadata_all) is called when pass doesn't
  make progress

v2: fix build

Signed-off-by: Marcin Ślusarz <marcin.slusarz@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12324>
2021-08-11 11:23:30 +00:00
Marcin Ślusarz 70723f278f microsoft/clc: preserve only valid metadata in clc_lower_printf_base
Signed-off-by: Marcin Ślusarz <marcin.slusarz@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/12324>
2021-08-11 11:23:30 +00:00
Jesse Natalie 4c5f1ef3ca microsoft/clc: Add a test for compiling a kernel with a read-write image
Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10163>
2021-07-29 14:54:48 -07:00
Jesse Natalie 6126dd29a4 microsoft/compiler: Implement texture loads from UAVs
This comes to the backend as image_load rather than txf but the
DXIL intrinsic is the same at the end of the day.

Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10163>
2021-07-29 14:54:31 -07:00
Caio Marcelo de Oliveira Filho 8af6766062 nir: Move workgroup_size and workgroup_variable_size into common shader_info
Move it out the "cs" sub-struct, since these will be used for other
shader stages in the future.

Reviewed-by: Kenneth Graunke <kenneth@whitecape.org>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/11225>
2021-06-08 09:23:55 -07:00
Caio Marcelo de Oliveira Filho c8a7bd0dc8 nir: Rename WORK_GROUP (and similar) to WORKGROUP
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>
2021-06-07 22:34:42 +00:00
Caio Marcelo de Oliveira Filho a71a780598 nir: Rename nir_intrinsic_load_local_group_size to nir_intrinsic_load_workgroup_size
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>
2021-06-07 22:34:42 +00:00
Caio Marcelo de Oliveira Filho 430d2206da compiler: Rename local_size to workgroup_size
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>
2021-06-07 22:34:42 +00:00
Jesse Natalie fa677c8644 nir_lower_readonly_images_to_tex: Support non-CL semantics
For non-CL, intrinsic access isn't set, because the image type doesn't
have access qualifier. Instead, the access qualifier is set on the variable.

So, add a mode to this pass which can chase back to the variable in addition
to the intrinsic access. Also, update the variable type and the deref chain
types so everything is consistent, that the tex is accessing a sampler. Note
we can't do this for CL, because void-typed samplers don't exist.

Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10356>
2021-04-23 23:16:15 +00:00
Jesse Natalie 29c9731400 nir: Rename nir_lower_cl_images_to_tex, replace 'cl' with 'readonly'
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10356>
2021-04-23 23:16:15 +00:00
Jesse Natalie 53786d2eb2 microsoft/clc: Fix undeclared function warning
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-By: Bill Kristiansen <billkris@microsoft.com>
Cc: mesa-stable@lists.freedesktop.org
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10331>
2021-04-20 00:28:34 +00:00
Jesse Natalie d641adca6a microsoft/clc: Fix MSVC unreferenced variable warnings
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Reviewed-By: Bill Kristiansen <billkris@microsoft.com>
Cc: mesa-stable@lists.freedesktop.org
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10331>
2021-04-20 00:28:34 +00:00
Jesse Natalie 8c140006e2 microsoft/clc: If local size isn't specified either in the shader or at runtime, set it to (1,1,1)
Otherwise we can end up in situations like having divide-by-zero. If the optimization is smart enough
that we end up with a *constant* divide-by-zero, then the DXIL validator will fail to sign, which
can trigger fatal errors with CLOn12.

We want to run an initial translation of all kernels during program build, but at that point we don't
know the local size to be able to specify it through kernel specialization data.

v2: Metadata output of 0 is used to indicate that the size wasn't explicitly specific. Copy the
    size to the metadata before overriding it to (1,1,1). If conf was explicitly specified,
    update the metadata again (though nobody should be paying attention to it).

Closes: https://github.com/microsoft/OpenCLOn12/issues/20
Closes: https://github.com/darktable-org/darktable/issues/8700
Reviewed-By: Bill Kristiansen <billkris@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10303>
2021-04-19 15:38:57 +00:00
Jesse Natalie debdc81ccf microsoft/compiler: Support arrays of UBOs
Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10149>
2021-04-16 17:08:17 +00:00
Jesse Natalie 8cc3100226 microsoft/clc: Update unit test to always use COMMON state for buffers
New Windows versions have new debug validation warning that for resources that
aren't actually UPLOAD/READBACK (which these aren't, thanks to the
GetCustomHeapProperties call), initial state that's not COMMON doesn't actually
do anything, which causes these tests to all fail because they verify that
they don't produce debug layer messages.

Reviewed-By: Bill Kristiansen <billkris@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10190>
2021-04-12 18:50:20 +00:00
Lionel Landwerlin ec6d9b7fb2 microsoft: fixup clc_log() define
The local msg variable shadows one of the argument of
SPIRVMessageConsumer making the error message "(null)".

Signed-off-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10133>
2021-04-09 15:29:39 +03:00
Bas Nieuwenhuizen 580f1ac473 nir: Extract shader_info->cs.shared_size out of union.
It is valid for all stages, just 0 for most of them. In particular
mesh/task shaders might be using it.

Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10094>
2021-04-08 14:39:28 +00:00
Dave Airlie 01dfd65a2d nir: port fp16 casting code from dxil
This moves the dxil pass to common code and makes dxil
use the new code.

Acked-by: Adam Jackson <ajax@redhat.com>
Reviewed-by: Roland Scheidegger <sroland@vmware.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9643>
2021-03-22 12:16:59 +10:00
Jason Ekstrand e20e85f01e nir: Make nir_ssa_def_rewrite_uses_after take an SSA value
This replaces the new_src parameter of nir_ssa_def_rewrite_uses_after()
with an SSA def, and rewrites all the users as needed.

Acked-by: Alyssa Rosenzweig <alyssa@collabora.com>
Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Reviewed-by: Eric Anholt <eric@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9383>
2021-03-08 16:59:55 +00:00
Jason Ekstrand 117668b811 nir: Make nir_ssa_def_rewrite_uses take an SSA value
This commit replaces the new_src parameter of nir_ssa_def_rewrite_uses()
with an SSA def, removes nir_ssa_def_rewrite_uses_ssa(), and rewrites
all the users as needed.

Reviewed-by: Rhys Perry <pendingchaos02@gmail.com>
Acked-by: Alyssa Rosenzweig <alyssa@collabora.com>
Acked-By: Mike Blumenkrantz <michael.blumenkrantz@gmail.com>
Reviewed-by: Eric Anholt <eric@anholt.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9383>
2021-03-08 16:59:55 +00:00
Jesse Natalie 8a3dbf1ca6 microsoft/clc: Add a test with an unused kernel arg
Reviewed-by: Bill Kristiansen <billkris@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9160>
2021-02-19 16:36:46 +00:00