Commit graph

21 commits

Author SHA1 Message Date
Boris Brezillon
afb64e10c1 microsoft/compiler: Move d3d12_fix_io_uint_type() to dxil_nir.c
We currently have two implementations of the same logic. Let's pick
the d3d12 one, move it to dxil_nir.c and let nir_to_dxil() call it
when appropriate.

Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/17039>
2022-06-15 22:04:00 +00:00
Louis-Francis Ratté-Boulianne
b0b9b50077 d3d12: Move d3d12_nir_lower_vs_vertex_conversion() to a common place
So we can re-use it in dozen.

Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/15955>
2022-06-14 17:42:51 +00:00
Boris Brezillon
b12417a2c7 microsoft/compiler: Lower fquantize2f16
As far as I can't tell, there's no native operation doing this
equivalent of fquantize2f16. Let's lower this operation to

   if (val < MIN_FLOAT16)
      return -INFINITY;
   else if (val > MAX_FLOAT16)
      return -INFINITY;
   else if (fabs(val) < SMALLER_NORMALIZED_FLOAT16)
      return 0;
   else
      return val;

which matches the definition of OpQuantizeToF16:

"
If Value is an infinity, the result is the same infinity.
If Value is a NaN, the result is a NaN, but not necessarily the same NaN.
If Value is positive with a magnitude too large to represent as a 16-bit
floating-point value, the result is positive infinity. If Value is negative
with a magnitude too large to represent as a 16-bit floating-point value,
the result is negative infinity. If the magnitude of Value is too small to
represent as a normalized 16-bit floating-point value, the result may be
either +0 or -0.
"

Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/16959>
2022-06-10 07:45:05 +00:00
Jesse Natalie
cc805aef69 d3d12: Properly set HS input control point count
Looks like some hardware needs this info in the shader to match the
topology. Since there's no spot in the shader info for it, we're
currently using the array size of the TCS input vars to store it.

Cc: mesa-stable
Reviewed-by: Paul Dodzweit <paul.dodzweit@amd.com>
Tested-by: Paul Dodzweit <paul.dodzweit@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/16920>
2022-06-09 17:08:18 +00:00
Enrico Galli
0badd0547d microsoft/spirv_to_dxil: Add pass to lower dynamic accesses on ubo[1]
Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14988>
2022-04-15 00:19:55 +00:00
Jesse Natalie
97b6ea71a0 microsoft/compiler: Add a pass for hull and domain shaders to shrink tess level vars
DXIL validation will complain if the tess factor signature entries have the
wrong number of components for the shader's domain. Make sure that both
hull and domain shaders have the right number, and drop loads and stores
from the removed components.

Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Reviewed-by: Bill Kristiansen <billkris@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14399>
2022-01-26 01:31:35 +00:00
Jesse Natalie
f511354a64 microsoft/compiler: Split hull (tess ctrl) shaders into main and patch constant funcs
Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Reviewed-by: Bill Kristiansen <billkris@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14399>
2022-01-26 01:31:35 +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
64991d44a8 microsoft/compiler: Load synthesized sysvals via lowered io
Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14175>
2021-12-20 08:20:59 -08:00
Jesse Natalie
11eb03c44e microsoft/compiler: Remove algebaric pass for inot
Reviewed-by: Alyssa Rosenzweig <alyssa@collabora.com>
Reviewed-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/14140>
2021-12-10 23:23:29 +00: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
Michael Tang
36aa1a2959 gallium/d3d12: move d3d12_lower_bool_input to microsoft/compiler
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Reviewed-by: Boris Brezillon <boris.brezillon@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13262>
2021-10-08 17:21:33 +00:00
Enrico Galli
e2e22720dc d3d12, microsoft/compiler: Moving driver_location allocation to compiler
Acked-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10989>
2021-06-24 20:05:12 +00:00
Jesse Natalie
46bc7cf678 microsoft/compiler: Rewrite sampler splitting pass to be smarter and handle derefs
Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10298>
2021-04-27 23:13:19 +00:00
Enrico Galli
3a3ab2f8ee microsoft/compiler: zero out unused WebGPU system values
Reviewed-by: Jesse Natalie <jenatali@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10439>
2021-04-27 01:53:16 +00:00
Jesse Natalie
5578fdd2c4 microsoft/compiler: Add a lowering pass to emit double [un]pack instructions
Whenever we have an ALU op that's operating on a double, we'll unpack
it as an integer, then repack it as a float. When we have an ALU op that
returns a double, we'll unpack it as a double, then repack it as an integer.

Then, simple algebraic opts will remove any redundant unpack/repack ops,
so we should be left with constructing and deconstructing doubles using
the right operations.

Reviewed-by: Enrico Galli <enrico.galli@intel.com>
Reviewed-by: Michael Tang <tangm@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/10063>
2021-04-09 01:54:33 +00:00
Jesse Natalie
4941decc58 microsoft/compiler: Enable dxil_nir.h to be included from C++
Reviewed-by: Bill Kristiansen <billkris@microsoft.com>
Acked-by: Michael Tang <tangm@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9846>
2021-04-02 17:16:48 +00:00
Jesse Natalie
65e1bee99c microsoft/compiler: Add a lowering pass to split clip/cull distance compact arrays
As the pass's comment explains, NIR compact arrays are a better match
to represent the GLSL float[] design of clip/cull distance arrays, compared
to the float4[2] approach. This pass adjusts the variables to more closely
match what DXIL signatures would look like for that representation.

Reviewed-by: Bill Kristiansen <billkris@microsoft.com>
Acked-by: Michael Tang <tangm@microsoft.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/9846>
2021-04-02 17:16:48 +00:00
Jesse Natalie
ff05da7f8d microsoft: Add CLC frontend and kernel/compute support to DXIL converter
This adds a standalone library which can convert through the pipeline of
OpenCL C -> SPIR -> SPIR-V -> NIR -> DXIL. It can add in the libclc
implementations of various library functions in the NIR phase, and
also massages the NIR to shift it more towards graphics-style compute.

This is leveraged by the out-of-tree OpenCLOn12 runtime
(https://github.com/microsoft/OpenCLOn12).

This is the combination of a lot of commits from our development branch,
containing code by several authors.

Co-authored-by: Boris Brezillon <boris.brezillon@collabora.com>
Co-authored-by: Daniel Stone <daniels@collabora.com>
Co-authored-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Acked-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7565>
2020-11-18 04:05:37 +00:00
Vinson Lee
50175716d2 microsoft/compiler: Add dxil_nir_lower_16bit_conv prototype.
Fix missing-prototypes build error.

src/microsoft/compiler/dxil_nir_algebraic.c:5761:1: error: no previous prototype for ‘dxil_nir_lower_16bit_conv’ [-Werror=missing-prototypes]
 5761 | dxil_nir_lower_16bit_conv(nir_shader *shader)
      | ^~~~~~~~~~~~~~~~~~~~~~~~~

Fixes: b9c61379ab ("microsoft/compiler: translate nir to dxil")
Signed-off-by: Vinson Lee <vlee@freedesktop.org>
Reviewed-by: Erik Faye-Lund <erik.faye-lund@collabora.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7559>
2020-11-11 12:09:52 -08:00
Erik Faye-Lund
b9c61379ab microsoft/compiler: translate nir to dxil
Here's the code to emit DXIL code from NIR. It's big and bulky as-is,
and it needs to be split up a bit.

This is the combination of a lot of commits from our development branch,
containing code by several authors.

Co-authored-by: Bill Kristiansen <billkris@microsoft.com>
Co-authored-by: Boris Brezillon <boris.brezillon@collabora.com>
Co-authored-by: Daniel Stone <daniels@collabora.com>
Co-authored-by: Gert Wollny <gert.wollny@collabora.com>
Co-authored-by: Jesse Natalie <jenatali@microsoft.com>
Co-authored-by: Louis-Francis Ratté-Boulianne <lfrb@collabora.com>
Acked-by: Jason Ekstrand <jason@jlekstrand.net>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7477>
2020-11-10 15:37:07 +00:00