This change implements pointers and references as described by the WGSL
specification change in https://github.com/gpuweb/gpuweb/pull/1569.
reader/spirv:
* Now emits address-of `&expr` and indirection `*expr` operators as
needed.
* As an identifier may now resolve to a pointer or reference type
depending on whether the declaration is a `var`, `let` or
parameter, `Function::identifier_values_` has been changed from
an ID set to an ID -> Type* map.
resolver:
* Now correctly resolves all expressions to either a value type,
reference type or pointer type.
* Validates pointer / reference rules on assignment, `var` and `let`
construction, and usage.
* Handles the address-of and indirection operators.
* No longer does any implicit loads of pointer types.
* Storage class validation is still TODO (crbug.com/tint/809)
writer/spirv:
* Correctly handles variables and expressions of pointer and
reference types, emitting OpLoads where necessary.
test:
* Lots of new test cases
Fixed: tint:727
Change-Id: I77d3281590e35e5a3122f5b74cdeb71a6fe51f74
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50740
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Currently we use offset to calculate the index of the queries in
timestamp compute shader, which is incorrect. The offset is the buffer
offset where we start to write the query results, and has nothing to
do with query index. In the query availability detection, the query
index should be based on the parameter firstQuery.
Add new test for resolving a timestamp query twice to the same
destination buffer with potentially overlapping ranges.
Bug: dawn:434
Change-Id: I2b5c5b192cf5d987ac48187e8240a25937957f51
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50760
Reviewed-by: Austin Eng <enga@chromium.org>
Commit-Queue: Hao Li <hao.x.li@intel.com>
Don't construct these with undefined values
Change-Id: I6225d9be0973ebc1a8594526aa32ec6775b5e865
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51300
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Rename:
* type() to StoreType()
* storage_class() to StorageClass()
Move away from snake_case methods in the semantic namespace.
Try to avoid generic 'type()' method names.
Also add an assertion to detect doubly nested references (these are
invalid).
Bug: tint:727
Change-Id: I975a3f1e5fbed7947cc2fc156fee892b282c63de
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51220
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
This permits enabling the OpenGL and OpenGL ES backends independently.
This change also enables the OpenGL ES backend on Windows.
This will cause the end2end tests to run on OpenGL ES on the
GPU-less bots, via SwANGLE.
Bug: dawn:580
Change-Id: I43d514b18862d176610b95e97013a67723ddac20
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50881
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Commit-Queue: Stephen White <senorblanco@chromium.org>
Adds texture_external to the inspector, allowing us to recognize the
type and return provide binding information. Includes a basic test.
Bug: Dawn:728
Change-Id: Ib0f39998359dc22a530ad222141229f9ba30552f
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51161
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Commit-Queue: Brandon Jones <brandon1.jones@intel.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Once the resolver correctly handles address-of and indirections,
the logic around these will be different enough that a parameterized
test no longer makes sense.
Bug: tint:727
Change-Id: I79c138b12bd6f49d9dbee02c62e594a9b1b4a2db
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51185
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: David Neto <dneto@google.com>
Add AddressOf() and Deref()
Add overloads of Expr() that take a source
Change WrapInStatement() to create a `let`. Unlike `var`, `let` can be
used to hold pointers.
Bug: tint:727
Change-Id: Ib2cd7ab7a7056862e064943dea04387f7e466212
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51183
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Force the use_tint_generator toggle so that Tint is used as much as
possible in fuzzers to get coverage of the relevant code.
This will also fix a recurring fuzzer issue that required a code
transform that only exists in the Tint path: bindings must be renumbered
otherwise we hit an integer overflow in Swiftshader's descriptor set
writes.
Bug: chromium:1209602
Change-Id: Ie868fb07306f96a8ecf42a687313e0a466041a45
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/51240
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>
Auto-Submit: Corentin Wallez <cwallez@chromium.org>
Commit-Queue: Ryan Harrison <rharrison@chromium.org>
Use this to check that type and expr are both not nullptr.
Cleans up the random mix of checking `expr`, `type` or both.
Change-Id: I84eb4a16e0dc22e18d6868f62c0314effe9858a2
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51184
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Currently unused, but will be soon.
Also add String() methods on all the types ifndef DEBUG.
Again unused, but these are helpful for debugging type related bugs.
Bug: tint:727
Change-Id: I7598fd7f68f4819ef4b60dbfc08322fd004d9152
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51180
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
This is what it is (currently) called in the spec
Bug: tint:727
Change-Id: Ie24f42499ed20c0c45ef4e9474bc6bb6a19bfa36
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51181
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
The change renames Texture2DCopySplit and copies2D to
TextureCopySubresource and copySubresources respectively.
Because they are not used for 2D only.
I didn't change Texture2DCopySplit to TextureCopySplit in order
to reflect its meaning and distinguish it from TextureCopySplits.
TextureCopySubresource is a collection of copy regions for either a
single layer of a 1D/2D texture or all depth slices on the same
mip level of a 3D texture (They are exactly what subresources are).
It also renames function ComputeTextureCopySplit to
ComputeTextureCopySubresource, and a couple similar renaming.
Bug: dawn:547
Change-Id: I17f8b349e209af0ed1ccaee4634be1e8235a63b3
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50920
Commit-Queue: Corentin Wallez <cwallez@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Use the device maximum supported shader model for DXC compiler.
Bug: None
Change-Id: I05f66d499405a3512a43daca35bb20be84f1f6c5
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/51080
Commit-Queue: Corentin Wallez <cwallez@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>
Reviewed-by: Rafael Cintron <rafael.cintron@microsoft.com>
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Semantic information about block statements the resolver would
temporarily create while resolving is now exposed in a
sem::BlockStatement class.
In the process, semantic information about statements in general is
overhauled so that a statement has a reference to its parent
statement, regardless of whether this is a block.
Bug: tint:799
Bug: tint:800
Change-Id: I8771511c5274ea74741b8c86f0f55cbc39810888
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50904
Commit-Queue: Alastair Donaldson <allydonaldson@googlemail.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Also validate that workgroup_size is only applied to compute stages,
and not duplicated.
Fixed: tint:703
Change-Id: I02f4ddea305cad25ee0a99e13dc9e7fd1d5dc3ea
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51120
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Auto-Submit: James Price <jrprice@google.com>
Change-Id: Iaa098d7851ffaf01aac384fe623c94809948fe2e
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51182
Commit-Queue: Ben Clayton <bclayton@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
In preparation for implementing
https://github.com/gpuweb/gpuweb/issues/1604, this change removes the
sem::AccessControl node. Instead, the ast::AccessControl::Access enum is
now on the sem::StorageTexture class, as well as on sem::Variable. For
sem::Variable, the field is set when the variable's type is either a
storage buffer or a storage texture.
Bug: tint:802
Change-Id: Id479af36b401d067b015027923f4e715f5f69f25
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51020
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
The backends now use sem::Variable::ConstantId() instead, since the
AST does not have the correct ID when the parameterless version of the
override attribute is used.
Change-Id: I816295c8b2e4baa1671c9d042920bf001b1b4336
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50845
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
Prove this error is caught by SPIR-V validation.
This requires a recent version of SPIRV-Tools.
Fixed: tint:793
Change-Id: I30977de7c4d1c291ab9ea13df80189426a842521
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51040
Auto-Submit: David Neto <dneto@google.com>
Commit-Queue: Alan Baker <alanbaker@google.com>
Reviewed-by: Alan Baker <alanbaker@google.com>
The Resolver, Inspector, and backends already handle this form of the
attribute correctly, so this is straightforward.
Fixed: tint:755
Change-Id: I8c394f1e58d64827d8a53b81402e30a472ca5441
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50844
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
This will be needed by Dawn to handle overriding pipeline constants by
name in the API.
Bug: tint:755
Change-Id: I857d8ef086b1964dac8d5cfff5c78d794234e8a8
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50843
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Generator::Generate() no longer takes an argument.
Change-Id: Ibee65ffd6672fb1507aec33c7690c5b05ce51fac
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50960
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: James Price <jrprice@google.com>
This enables the backends to emit overridable constants that had no ID
specified in the attribute.
Bug: tint:755
Change-Id: I86587205e065715257f546b546e792a5262562e8
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50842
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Keep track of any constant IDs specified in the shader, and then
allocate IDs for the remaining constants when creating the semantic
info.
Bug: tint:755
Change-Id: I6a76b1193cac459b62582cde7469b092dde51d5d
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50841
Commit-Queue: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Nothing currently generates these.
Resolver currently doesn't know how to handle these.
Backends currently stubbed enough to build without warnings.
Bug: tint:727
Change-Id: I4b9863ae098b903b51a63c36c10bc6e390efbbb3
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50746
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Reviewed-by: James Price <jrprice@google.com>
There's now one remainin use of the semantic info in the WGSL writer - the transformation of [[offset]] into padded fields.
This does not belong in the WGSL writer, but instead part of the transform::Wgsl sanitizer.
Bug: tint:798
Change-Id: I95ba11f022c41150cc12de84a4085cd7d42019fe
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50822
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Reviewed-by: James Price <jrprice@google.com>
The previous matmul is using 1-dimensional shared array. This PR
adds the 2-dimensional shared array. The perf result shows that:
1. For 1-dimensional shared array, enabe-robustness wil bring almost 2x
regression both for matmulFloat and matmulVec4 on Intel CFL.
2. For 2-dimensional shared array, enabe-robustness will bring almost 2x
regression on matmulFloat. But have little impact on matmulVec4 on Intel
CFL.
Tested on Intel_R_UHD_Graphics_630.
shader enable robustness disable robustness
MatMulFloatOneDimSharedArray 5383 us 3105 us
MatMulFloatTwoDimSharedArray 4788 us 2608 us
MatMulVec4OneDimSharedArray 3070 us 1743 us
MatMulVec4TwoDimSharedArray 1840 us 1802 us
Bug: dawn:594
Change-Id: Ia29a78cf70649ef8d3ba8476db1ad4d6ded80840
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50481
Reviewed-by: Jiawei Shao <jiawei.shao@intel.com>
Reviewed-by: Austin Eng <enga@chromium.org>
Commit-Queue: Jiajia Qin <jiajia.qin@intel.com>
To declare a local variable, we write `var name : type`, not `var<function> name : type`.
This change fixes all the places where we were feeding StorageClass::kFunction into variable declarations.
Note that the resolved, semantic variable correctly infers the `kFunction` StorageClass.
Change-Id: I6221fabae1de0435044f29b9a91808421d5cace6
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50821
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Commit-Queue: David Neto <dneto@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Added in https://dawn-review.googlesource.com/c/tint/+/50742
Bindings may not collide except on the D3D12 backend where we use
the register offset as the binding number. Offsets may alias if
they are for different HLSL register types (buffer, texture, etc.)
Bug: tint:797
Change-Id: I020e3b5cd1d35fdf5678b587225507e0cf24f5b3
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50801
Commit-Queue: Austin Eng <enga@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: Austin Eng <enga@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
Add a new parameter to BindingRemapper::Remappings that allows resulting binding points to collide.
When enabled, the output of the transform contains two or more module-scoped variables with the same binding point, used by the same entry point, then these variables will be decorated with an internal decoration to disable validation for the collision.
This is to work around collisions generated for the HLSL backend where the variables actually exist in different register classes, which is permitted by D3D12.
The transform will only generate these decorations if it needs to.
Fixed: tint:797
Change-Id: Id8a87523801bd0cd0dd54227ebabd4299bc20c27
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50742
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
An [[internal]] decoration that specifically disables certain validation checks.
Begin with a single kFunctionHasNoBody mode.
Migrate the Resolver to using this instead of allowing any InternalDecoration to disable the checks for no-body.
Bug: tint:797
Change-Id: I213b9a6844a456775ede06d60e456d9f77a449d0
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50741
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Variable() is called for globals, locals and parameters. Much of the logic is the same.
Move all the common logic down into Variable(). This:
* Removes some yucky default parameters
* Adds type validation that was missing for globals (broken tests fixed)
* Gives me a single place to implement the Reference type wrapping
Bug: tint:727
Change-Id: I70f4a3603d7fa781da938508aa2a1bc80ec15d77
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50580
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Removes the texture_external overload for textureSample and replaces it
with a texture_external overload of textureSampleLevel to match merged
spec. Adds a transform that adds the implicit level parameter to
textureSampleLevel. Modifies unit tests to reflect change.
Bug: dawn:728
Change-Id: I2dbc9232b4343db1075be79fda0054231860f3b1
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50701
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Make all test programs valid.
Change-Id: Id7eb790519e3dec30a5b826f06585d277995b9b5
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50720
Auto-Submit: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
WGSL only has vector normalize, not scalar.
Scalar normalize is always 1.0, so return that directly.
Bug: tint:765
Change-Id: I08332c20922f5834f65284b9aaeb22995423171d
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50603
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: David Neto <dneto@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Emit AST equivalent to:
var foo: sampler;
instead of
var<handle> foo: sampler;
This is necessary to pass new checks in the resolver.
Change-Id: I5eca30ace5871086bc3c1783cde2225f1aee34ba
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50602
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: David Neto <dneto@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
But I liked the comment where it was.
I think the linter was wrong, but it doesn't like my argument.
Change-Id: Ie3bb6e2b4bd08fae6d72a076622b1e74f517f23c
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/50680
Auto-Submit: David Neto <dneto@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Certain backends performs transformations on the tint::Program in the ShaderModule.
These transformations may raise diagnostics.
Tint diagnostics generated from an underlying tint::Source::File need that File kept alive.
Change-Id: I32deceb4cedb419e914631d342884fc27972b5cb
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/50581
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: Austin Eng <enga@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>
