SPIR-V spec states:
> Each structure-type member that is a matrix or array-of-matrices must have be decorated with a MatrixStride Decoration
As already pointed out in https://dawn-review.googlesource.com/c/tint/+/59840, we were not handling arrays-of-matrices.
To do this correctly, we need the ast::StrideDecoration to be placed on the Matrix type, which is a much bigger change to support.
For now, chase the type, and error if we have a custom MatrixStride on an array of matrices, otherwise drop the decoration.
Bug: tint:1049
Fixed: tint:1088
Change-Id: Idcb75b3df88040836a03a14e0ca402ebee7be9a7
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60923
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: David Neto <dneto@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
This patch adds the missing input [[sample_mask]] to EntryPoint in
the Inspector and the related unittests to inspector_test.cc.
According to the latest WGSL SPEC [[sample_mask]] can be used as
both fragment shader input and output.
Bug: dawn:802, dawn:1032
Change-Id: I80640f26c7d771e35039fd7f3c941859729f26ce
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/61020
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Jiawei Shao <jiawei.shao@intel.com>
The refactored CanonicalizeEntryPointIO transform makes it much easier
to handle SPIR-V style IO as well, and doing this removes a lot of
duplicated code. Remove all of the SPIR-V transform code for shader IO
and vertex point size.
Bug: tint:920
Change-Id: Id1b97517619b4d2fd09b45d5aee848259f3dfa77
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60840
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This option is passed through to the CanonicalizeEntryPointIO
transform, which adds it to the set of builtin shader outputs for all
vertex shaders in the module.
Bug: tint:1000
Change-Id: Ibba4adde2c468b11ebfd7012fcb42ee48aad04e4
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60522
Auto-Submit: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This patch adds all the fragment builtin inputs (position, front_face,
sample_index) to EntryPoint for the validation on the total number of
fragment inputs. According to Vulkan SPEC: "All variables in both the
built-in interface block and the user-defined variable interface count
against these limits".
BUG=dawn:802
Change-Id: I8a8503c1a33646b50f010c6b6e38d74de9a40ff5
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59421
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Jiawei Shao <jiawei.shao@intel.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
This is a major reworking of this transform. The old transform code
was getting unwieldy, with part of the complication coming from the
handling of multiple return statements. By generating a wrapper
function instead, we can avoid a lot of this complexity.
The original entry point function is stripped of all shader IO
attributes (as well as `stage` and `workgroup_size`), but the body is
left unmodified. A new entry point wrapper function is introduced
which calls the original function, packing/unpacking the shader inputs
as necessary, and propagates the result to the corresponding shader
outputs.
The new code has been refactored to use a state object with the
different parts of the transform split into separate functions, which
makes it much more manageable.
Fixed: tint:1076
Bug: tint:920
Change-Id: I3490a0ea7a3509a4e198ce730e476516649d8d96
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60521
Auto-Submit: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Fix the Renamer to preserve builtin structure member names.
Fix the HLSL writer to emit the modf / frexp result type even if there is no private / function storage usage of the types.
Fixed: chromium:1236161
Change-Id: I93b9d92980682f9a9cb090d07b04e4c3f6a2f705
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60922
Reviewed-by: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
We're about to go into a Chrome Origin Trial, and crbug.com/tint is not the correct place to file Chrome related bugs.
Move this message to the tint executable, which is code not used by chrome.
Change-Id: Ib33630164fa5d7bec3c581e5eca6ff188921eef8
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60562
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>
FXC errors on these, and they are undefined behavior in WGSL.
Bug: tint:1083
Change-Id: I7643fdc6991f8729f274535b603b761398412398
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60500
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Make ParserImp::const_literal() bubble up any error by the tokenizer.
These were being ignored.
Also:
* Detect and report significand too large
* Detect and report missing exponent
* Fix invalid mantissa overflow detection for fractional trailing zeroes
* Fix zero with non-zero exponent triggering an assert, and instead,
make the result zero (added tests for this).
Bug: chromium:1235132
Bug: tint:77
Change-Id: I364a4c944121a2c55ff3161de1bb50126c8a5526
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60680
Reviewed-by: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
FXC does not support fallthrough case statements (DXC does).
Fixed: tint:1082
Change-Id: I82e1add5455e438056259f773f34bf9db05970b4
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60480
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Instead of the CWD.
Can improve performance of the test-runner, as emitting a lot of short-lived files in the source tree can waist a lot of cycles triggering IDE file monitoring logic.
Change-Id: I25de15af02ab816fff5d8a079fda901883793478
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60342
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
https://dawn-review.googlesource.com/c/tint/+/60213 special cased ignore() to work around tint:1046.
This fix produced bad output for structures when they are fully decomposed into ByteAddressBuffers, as the final HLSL references a structure that no longer exists.
Fixes CTS tests, and tint->dawn roll.
Change-Id: If6eab083c5f0bcca4a90c582df255b77e97a8e9f
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60347
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Spread the array zeroing across as many workgroup invocations as possible.
Bug: tint:910
Change-Id: I1cb5a6aaafd2a0a4093ea3b9797c173378bc5605
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60203
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
This reverts commit e5dbe24e94.
Reason for revert: Makes the Tint-Dawn roll fails because of
MSL compilation errors on as_type<uint>(-2147483648):
as_type cast from 'long' to 'uint' (aka 'unsigned int') is not allowed
as_type<uint>(-2147483647) compiles fine, so this is most
likely because the MSL compiler types the literal as a long
(since without the - it is larger than the max int32).
Original change's description:
> MSL writer: make signed int overflow defined behaviour
>
> Bug: tint:124
> Change-Id: Icf545b633d6390ceb7f639e80111390005e311a1
> Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60100
> Kokoro: Kokoro <noreply+kokoro@google.com>
> Commit-Queue: Antonio Maiorano <amaiorano@google.com>
> Reviewed-by: David Neto <dneto@google.com>
TBR=dneto@google.com,bclayton@google.com,jrprice@google.com,amaiorano@google.com,noreply+kokoro@google.com,tint-scoped@luci-project-accounts.iam.gserviceaccount.com
Change-Id: I3e3384a9185013bb141a1b7b9b22bad8571bbc50
No-Presubmit: true
No-Tree-Checks: true
No-Try: true
Bug: tint:124
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60345
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
Auto-Submit: Corentin Wallez <cwallez@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
For loops only support assignments or function calls for the continuing statement.
Fixed: tint:1064
Change-Id: I07065b2119e7b9f97ca7e46b1464fd72333ca429
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60212
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
It wasn't unwrapping the reference before type checking
Change-Id: I4bfc038c468c32c2a164bbcbef0a97a3e385d5ba
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60210
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
- function scope variable store type must be constructible
- add IsConstructible() to sem::atomic
Bug: tint:1069
Change-Id: Ib0616b486ecf278dbdd99640dc4ede7f3007feb8
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60120
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Sarah Mashayekhi <sarahmashay@google.com>
Auto-Submit: Sarah Mashayekhi <sarahmashay@google.com>
Actually call Expression() on the workgroup sizes.
This generates the semantic information for the expressions that would otherwise be missing.
Bug: tint:910
Change-Id: I9d7f9d6b029165dfb3bd1e0bf7ce86c0a71dd4d5
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60205
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
The second parameter must not be a u32.
Fixed: tint:1078
Bug: tint:1079
Change-Id: Id7a9cd881c4fec0f262931c2e4c263310e59c25d
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60204
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
- 'location' attribute must only be applied to declarations of numeric scalar or numeric vector type
- 'location' attribute is not valid for compute shader
- locations must not overlap
Bug: tint:1035
Change-Id: I0ba301996f390c8206192d2f81e787e0eac0aa6a
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59760
Kokoro: Kokoro <noreply+kokoro@google.com>
Auto-Submit: Sarah Mashayekhi <sarahmashay@google.com>
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
When building a vector via tint::writer::AppendVector, and the
vector argument is already a vector constructor, expand that
vector constructor into its components only when those components
are all scalars. This avoids a type breakage which can occur with cases
like this:
vector argument is:
vec2<i32>(vec2<u32>(0u,1u))
scalar argument is:
2
Before this fix, the result was:
vec2<i32>(0u, 1u, 2);
But should be this instead:
vec3<i32>(vec2<u32>(0u,1u),2)
This was noticed in SPIR-V writer output when forming a coordinate
vector from a an unsigned WGSL coordinate vector with a signed array
vector.
Fixed: tint:1048
Change-Id: Id46665739cc23da0ca58b9baabf7b4531b86350b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/60040
Auto-Submit: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: David Neto <dneto@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This code was implicitly assuming that all resources it was looking
for would be directly referenced at the intrinsic callsite, and not
passed via function parameters.
This was causing a crash in more complex cases.
The inspector code has been updated to handle cases where the
resources are not being directly referenced.
Unneeded calls to GenerateSamplerTargets() are removed.
Utility function GetOriginatingResources() is added to handle walking up
call sites to resolve resources.
Text shader based test runner is added to the Inspector tests to make
expressing complex tests easier.
BUG=tint:967
Change-Id: I2ecb6d57c518003da59f38b261bae4d62ce7e6ac
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59340
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ryan Harrison <rharrison@chromium.org>
Auto-Submit: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
The continuing block can exit the loop in very constrained ways:
When a break statement is placed such that it would exit from a loop’s
§ 7.3.8 Continuing Statement, then:
- The break statement must appear as either:
- The only statement in the if clause of an if statement that has:
- no else clause or an empty else clause
- no elseif clauses
- The only statement in the else clause of an if statement that has an
empty if clause and no elseif clauses.
- That if statement must appear last in the continuing clause.
By design, this allows a lossless round-trip from SPIR-V to WGSL and
back to SPIR-V. But that requires this special case construct in WGSL
to be translated to an OpBranchConditional with one target being
the loop's megre block (which is where 'break' branches to), and the
other targets the loop header (which is the loop backedge). That
OpBranchConditional takes the place of the normal case of an
unconditional backedge.
Avoids errors like this:
continue construct with the continue target X is not
post dominated by the back-edge block Y
Fixed: 1034
Change-Id: If472a179380b8d77af746a3cd8e279c8a5e56b37
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59800
Auto-Submit: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: David Neto <dneto@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Add `transform::DecomposeStridedMatrix`, which replaces matrix members of storage or uniform buffer structures, that have a [[stride]] decoration, into an array
of N column vectors.
This is required to correctly handle `mat2x2` matrices in UBOs, as std140 rules will expect a default stride of 16 bytes, when in WGSL the default structure layout expects a stride of 8 bytes.
Bug: tint:1047
Change-Id: If5ca3c6ec087bbc1ac31a8d9a657b99bf34042a4
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59840
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Implemented for all readers and writers.
Cleaned up some verbose code in sem::Function and the Inspector in the
process.
Fixed: tint:1032
Change-Id: Ia6f2f59e6d2e511c89160b97be990e8b7c9828d9
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59664
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
This forbids let declarations from having handle types.
Change-Id: I6f7467b0fa3963711ec705e1a81bfdd2c550feee
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59801
Auto-Submit: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Pass a pointer argument instead of a reference argument.
Also handle the case where the argument is the result of
an OpCopyObject, which will generate a let-declaration of pointer
type.
Fixed: tint:1042
Change-Id: I25b1b7b95ade1b79130e51691194f32b3240e013
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59451
Auto-Submit: David Neto <dneto@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Use Private or Function storage instead of Workgroup storage.
Bug: tint:1043
Change-Id: I07055c30c8b89c9661c2a4913b0c03c3dff5a75b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59562
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
The DefInfo structure is used for remapping storage buffer types
as well as tracking special values like builtin variables.
In the latter case, don't take the defaulted storage class value
from the DefInfo initialization.
Fixed: tint:1040, tint:1043
Change-Id: I41ee364d76e632736d51f4474c97036bcc136c93
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59484
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
These work on scalar and vector of bool, and map to ast::BinaryOp::kOr
and kAnd.
Bug: tint:1043
Change-Id: I009edf8e43c21cb75ccfdcde1106ec177d2fe50e
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59561
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
These just print colors to stdout, and actually do not check anything.
They're pretty, but I'm fairly confident this works now, and doesn't actually give us anything other than spam output.
Disable.
Change-Id: I6144f154345ae115f0899ca7e274fdee8947c785
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59621
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
- skip dumping due to SPIR-V Tools validation error for Location
conflicts for IO variables duplicated on an OpEntryPoint interface
list (SPIRV-Tools issue #4403)
- Skip dumping for unit test for conversion of RelaxedPrecision
decoration.
Bug: tint:1043
Change-Id: I8ce3a7660d74b771e3fa62dc39e4e030e4ba6aa1
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59560
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
- each vertex shader must have builtin-position output.
- when testing non-vertex shader features, change to fragment shaders,
to avoid the need to have a builtin position output.
- Fix ordering of entry point declaration vs. OpName
Bug: tint:1043
Change-Id: Ic66383d42419a1ef893835ae71729ae27fc6e539
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59482
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
Add end to end tests
Fixed: tint:1026
Change-Id: I10813cbe6dc4f1bccddf9a8a29e3a249a364c051
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59663
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Implement these for all the writers.
SPIR-V reader not implemented (the old overloads weren't implemented either).
Deprecate the old overloads.
Fixed: tint:54
Change-Id: If66d26dbac3389ff604734f31b426abe47868b91
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59302
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
- clean up function decorations unit tests
- clean up interpolate and invariant validation and unittest
- add separate unit-tests for each shader stage input and output
- add [[builtin(position)]] tests
- add validation and test for:
structures with 'location' decorated members cannot be used as compute shaders input
Bug: tint:1007
Change-Id: I12e97e163b3a77bc76ce21faba241683eec5d917
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58942
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
This is a cleaner API, and the implementation doesn't have to know a bunch of information about all the derived types.
Change-Id: I96bebcb9f3ceda86fa34bd8e70961dee63fd7e13
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59301
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
`Size()` will be added which is the size of the type in bytes.
Change-Id: If997820d7859cd9d1bb0631d1b72150378e6a24b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59300
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
These are generated by the SPIR-V reader. Structures used by the
storage and uniform storage classes are handled separately, so we can
safely ignore this attribute for the other storage classes which are
not shared with the host.
Fixed: tint:1027
Change-Id: I8df4b41c687922a97f3b0ed97804cecc17d8997a
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59240
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Handles nested block comments.
Allow unterminated block comments at EOF, as it is not clear whether
WGSL will allow this or not.
Bug: tint:881
Change-Id: Ieae4e0073dab69f773adb32018a9bdaf4f352116
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59180
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Change-Id: I3de1e2437b9604378b8368494363e19070443670
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59201
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: James Price <jrprice@google.com>
Each of these may contain information specific to their kind.
Change-Id: Ic8ac808088132b7bc2e43da6ce46a06571e0fed5
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59200
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: James Price <jrprice@google.com>
Instead of concatenating to a string
Change-Id: Ib70bba2b0dbf9fcb04a232e0e78d3b576ab3ba88
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/59203
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
https://github.com/gpuweb/gpuweb/pull/1945 changes the SPIR-V mapping of this operator so that it now maps to OpFRem instead of OpFMod. Polyfill OpFMod with `x - y * floor(x / y)`
Also map the MSL output of this operator to use `fmod()`.
Behavior of this operator is now consistent across all backends.
Fixed: tint:945
Fixed: tint:977
Fixed: tint:1010
Change-Id: Iefa009b905989c55ace24e073ab0e261c7cf69b0
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58393
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Also fix cases of implicit conversions of bool to int when creating
sem::Array.
Bug: tint:917
Change-Id: I5392fb737efc410f039b4dbd96cffc5daa4fd3a2
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58783
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Each argument of a function call of pointer type must be one of:
- An address-of expression of a variable identifier expression
- A function parameter
Also added source location to duplicate struct member name unittest
Bug: tint:983
Change-Id: Ic5ab010b2ed76207a1d8d3ef9f66140ea95f7e72
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58480
Auto-Submit: Sarah Mashayekhi <sarahmashay@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@chromium.org>
Commit-Queue: Sarah Mashayekhi <sarahmashay@google.com>
Fuzzers like to generate silly long source, and formatting large spans of these can take considerable time.
Only format the diagnostic if it is going to be displayed.
Significantly speeds up some fuzzing tests, fixing some timeouts.
Also add a minor optimization to the formatter repeat() implementation.
Fixed: chromium:1230313
Change-Id: Ib1f6ac0b31010f86cb7f4e1432dc703ecbe52cb0
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58841
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: Ryan Harrison <rharrison@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Spvtools recently added support for the vkdebuginfo instructions,
causing roll failures since Tint build files didn't generate the
headers for them. Instead of adding more generator stuff in Tint's
BUILD.gn files, this commit removes everything in favor of directly
referencing spvtools' targets.
In follow-up work, the references to spvtools_internal_config will
be changed to a more narrowly scoped target so that spvtools'
warning suppressions don't get propagated to Tint.
Bug:chromium:1228274
Change-Id: I5e28c9cd978afd3cfbc941c09decf5a6e7e1554a
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58840
Commit-Queue: Corentin Wallez <cwallez@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
We have chosen not to implement this for origin trial, so just parse
it and produce an error in the validator.
Bug: tint:752
Change-Id: I40ebf0b4051fb0b8bab9d4c733e555b836320cd4
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58780
Auto-Submit: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Compute it in 2 dimensions, with a 0-valued y component,
then extract the x component of that result.
Fixed: tint:974
Change-Id: Ie23668d3403e68be14f34da9540f27f6f3c3aca2
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58782
Auto-Submit: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: James Price <jrprice@google.com>
This reflects the total size of all workgroup storage-class variables
referenced transitively by an entry point.
Bug: tint:919
Change-Id: If3a217fea5a875ac18db6de1579f004e368fbb7b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57740
Reviewed-by: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ken Rockot <rockot@google.com>
The only non-trivial case is SPIR-V, which generates OpSelect to
choose between 1 or 0.
Fixed: tint:997
Change-Id: Ifda7b3ec1e0a713843a2da7ed59c3449d4eec8bd
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58521
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: James Price <jrprice@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This is in preparation for handling the "invariant" decoration.
Bug: tint:972
Change-Id: I17465946932ab37a32dfd3c477525649ab622c6f
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58580
Auto-Submit: David Neto <dneto@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
The TypeConstructorExpression logic that tested for splats was not considering references. This led to broken emission for the SPIR-V and HLSL backends.
Fixed: tint:992
Change-Id: I9824b71f526997f91d380c09b459f4fd73065b19
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58397
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
We were not printing source information for errors involving call
statements.
Change-Id: I576b4d095162333d4392d6be590e827a49918d3d
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58520
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>
This is defensive. Without variable pointers capabilities, this is
definitely invalid, but not yet checked by the SPIRV-Tools validator.
Bug: tint:807
Change-Id: If9b0b19573b1ca14a1c55aa20c9d42784ec12568
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/56700
Auto-Submit: David Neto <dneto@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: David Neto <dneto@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Rework the clamping so that it unifies the logic for arrays, matricies
and vectors. Try to preserve constant signess, and only clamp the values
if they're possibly out of bounds.
Use ConstantValue() instead of scanning for ScalarConstantExpressions.
As ConstantValue() improves, so will the performance of robustness.
Change-Id: I013a67e15f43350d0a57bcd8ba9ae0c1bcb1eaec
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58280
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: David Neto <dneto@google.com>
Check they do not contain returns, discards
Check they do not directly contain continues, however a nested loop can have its own continue.
Bug: chromium:1229976
Change-Id: Ia3c4ac118ffdaa6cca6025366c19f9897718c930
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58384
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: David Neto <dneto@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Arrays can be extremely large, and having the load and store functions unroll the elements can make the complier explode.
Fixed: chromium:1229233
Change-Id: Ieb5654254e16f5ce724a205d21d954ef9a0cd053
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58382
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Reviewed-by: David Neto <dneto@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Have postfix_expression() use a loop instead of recursively calling itself.
Fixed chromium:1229669
Change-Id: Ied7684d00ba453a8b89ab0251d42e2a72169421f
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58381
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
NoPerspective interpolation maps to 'linear'
Centroid maps to 'centroid'
Sample maps to 'sample'
Otherwise, allow 'center' to be defaulted.
Fixed: tint:935
Change-Id: I0b040da0c57d2a363f9dc9474c1ac889e0fe2278
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/56840
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: David Neto <dneto@google.com>
Setting precision to `std::numeric_limits<float>::max_digits10` is valid
when using the `scientific` floatfield format when printing values.
However, we used `fixed` to make our floats more human-readable. This
change keeps the output in `fixed`, except if doing so loses precision,
in which case we fall back to `scientific`.
This fixes the rendering differences seen in the Babylon.js examples
(https://crbug.com/tint/944) between Dawn using Tint vs SPIRV-Cross, as
Tint's output was emitting values that had lost too much precision
(e.g. very small numbers being output as 0).
Bug: tint:944
Change-Id: I8deea23ad876825bbe390fc26907d4bbbd4b966e
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58321
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
For loop initializers and continuing statements do not have a BlockStatement as their parent.
Handle removal of these statements with a new Transform::RemoveStatement() helper
Fixed: tint:990
Change-Id: I24e7b18dcf71d3ef0a4d3ee68b9f68518e0eb5e8
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58063
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Some methods passed by pointer, others by reference. Standarize to pass-by-reference.
Also remove CloneWithStatementsAtStart().
CloneContext::InsertFront() is a better replacement.
Change-Id: Ibbf7caaa7a1b42c2d0a0cddaa3d6e76ca0e12a17
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58062
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Much like sem::Type, it greatly simplifies downstream logic if we can compare sem::Intrinsic pointers to know if they refer to the same intrinsic overload.
Change-Id: If236247cd3979bbde821d9294f304ab85ba4938e
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58061
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: David Neto <dneto@google.com>
And call these helpers instead of inlining complex statements.
Cleans up output, and helps prevent for-loops decaying to while loops.
Change-Id: I6ac31b18ce6c5fac0e54e982f7db3bb298f7edb2
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58060
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Reviewed-by: David Neto <dneto@google.com>
CloneContext::Replace(T* what, T* with) is bug-prone, as complex transforms may want to clone `what` multiple times, or not at all. In both cases, this will likely result in an ICE as either the replacement will be reachable multiple times, or not at all.
The CTS test: webgpu:shader,execution,robust_access:linear_memory:storageClass="storage";storageMode="read_write";access="read";atomic=true;baseType="i32"
Was triggering this brokenness with DecomposeMemoryAccess's use of CloneContext::Replace(T*, T*).
Switch the usage of CloneContext::Replace(T*, T*) to the new function form.
As std::function is copyable, it cannot hold a captured std::unique_ptr.
This prevented the Replace() lambdas from capturing the necessary `BufferAccess` data, as this held a `std::unique_ptr<Offset>`.
To fix this, use a `BlockAllocator` for Offsets, and use raw pointers instead.
Because the function passed to Replace() is called just before the node is cloned, insertion of new functions will occur just before the currently evaluated module-scope entity.
This allows us to remove the "insert_after" arguments to LoadFunc(), StoreFunc(), and AtomicFunc().
We can also kill the icky InsertGlobal() and TypeDeclOf() helpers.
Bug: tint:993
Change-Id: I60972bc13a2fa819a163ee2671f61e82d0e68d2a
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58222
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Before and after each transform of a transform::Manager.
This change makes the transforms use the Castable system so we can get the transform name.
Change-Id: I2cf9335960f1aca56f1d32c5ba0d104db4cc46c1
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/54581
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: James Price <jrprice@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Replace(T* what, T* with) is bug-prone, as more complex transforms may want to clone `what` multiple times, or not at all. In both cases, this will likely result in an ICE as either the replacement will be reachable multiple times, or not at all.
This is the cause of some of the CTS failures reported in crbug.com/tint/993.
Bug: tint:993
Change-Id: I880ece45faab0e7f07230a1b4436f4e9846edc84
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58221
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
This is a reland of fd5829e5ea
Original change's description:
> Validate storage class constraints
>
> As defined by https://gpuweb.github.io/gpuweb/wgsl/#storage-class-layout-constraints
>
> Bug: tint:643
> Change-Id: I9c78ba69a792a80c263a17b0a6e9b4810fdb7f30
> Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/56780
> Kokoro: Kokoro <noreply+kokoro@google.com>
> Commit-Queue: Antonio Maiorano <amaiorano@google.com>
> Reviewed-by: Ben Clayton <bclayton@google.com>
Bug: tint:643
Change-Id: I62036d615a062597339a9d130b7ccf49b5be26c7
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58120
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Clamp the `level` and `array_index` arguments to `textureLoad()` and `textureStore()`.
Also fix the off-by-one error for the coordinates.
See: https://github.com/gpuweb/gpuweb/pull/1906
Fixed: tint:748
Change-Id: Id7505578b632dcaf75b2a3a020fc0190c612972c
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57700
Kokoro: Kokoro <noreply+kokoro@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
It appears that I didn't do a great job cleaning up the removal of ShareableCloneable in https://dawn-review.googlesource.com/c/tint/+/51484.
Cloning nodes shouldn't return the same pointer. Remove bad comments.
Clean up leftover logic from CloneWithoutTransform().
Change-Id: Ibbc5f625c5978e9c11da59e7aa6197f39b6f0363
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/58220
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
This uses FXC compilation failure mitigation for _any_ vector index assignment that has a non-constant index. FXC can still fall over if the loop calls a function that performs the dynamic index.
Use some vector swizzle logic to avoid branches in the helper.
Fixed: tint:980
Change-Id: I2a759d88a7d884bc61b4631cf57feb4acc8178de
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57882
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
The UBO must have a stride that is a multiple of 16 bytes.
Note that this change was part of https://dawn-review.googlesource.com/c/tint/+/56780
but the CL was reverted because it broke Dawn. This CL relands part of
the change, and adds the macro TINT_EXPECTS_UBOS_TO_BE_MULTIPLE_OF_16 so
that Dawn can conditionally compile against it.
Bug: tint:984
Bug: tint:643
Change-Id: I303b3fe81ff97c4933c489736d5d5432a59ce9b7
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57921
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This was just echoing back the source that was provided to the
validation routine, and is no longer uses by any callers.
Change-Id: I55aeb7c41405ebbe1b21b9654b94607b651ff118
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57281
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This is a reland of the CL:
https://dawn-review.googlesource.com/c/tint/+/55402
Now that sanitizers are no longer exposed externally, the Inspector no
longer needs to handle this post-sanitizer world.
Bug: tint:697
Change-Id: Ic02ebb9c529aa132a238285bdd0d0df8686e219b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57104
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Make the sanitizer transforms internal, as the new generator API
automatically runs them.
Change-Id: Ia2674762328b5d91d8370b8c18c31693936e8566
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57102
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This change introduces sem::CompoundStatement, a new base class for
statements that can hold other statements.
sem::BlockStatements now derives from sem::CompoundStatement, and
this change introduces the following new CompoundStatements:
* `sem::IfStatement`
* `sem::ElseStatement`
* `sem::ForLoopStatement`
* `sem::LoopStatement`
* `sem::SwitchStatement`.
These new CompoundStatements are now inserted into the semantic
tree as now documented in `docs/compound_statements.md`.
The `sem::BlockStatement::FindFirstParent()` methods have been
moved down to `sem::Statement`.
The `Resolver::BlockScope()` method has been replaced with
`Resolver::Scope()` which now maintains the `current_statement_`,
`current_compound_statement_ ` and `current_block_`. This
simplifies statement nesting.
The most significant change in behavior is that statements now
always have a parent, so calling Block() on the initializer or
continuing of a for-loop statement will now return the
BlockStatement that holds the for-loop. Before this would
return nullptr.
Fixed: tint:979
Change-Id: I90e38fd719da2a281ed9210e975ab96171cb6842
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57707
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
Case statements cannot be used as regular statements. They must exist inside a switch.
Change-Id: Ica2e0573ad926c3e257f02c95397a8666dfcd547
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57705
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
This class is very large, so factoring it out into its own seperate
file.
Also renaming it to InspectorBuilder, in anticipation of adding tests
that don't build their state through explicitly calling into the
ProgramBuilder.
BUG=tint:967
Change-Id: I316458a969479200edf6962ab2008180f9a1e7e2
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57800
Auto-Submit: Ryan Harrison <rharrison@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ryan Harrison <rharrison@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
ConstantValueOf() obtains the constant value from the logic in resolver_constants.cc. This is better tested, and is the foundation of Tint's constant folder.
Change-Id: I42036f3ff4ab684b4864cd69856de1715b38d246
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57702
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Use the new semantic constant value information to significantly reduce the complex indexing logic emitted for UBO accesses.
This will dramatically reduce the number of `for` loops that are decayed to `while` loops.
Change-Id: I1b0adb5edde2b4ed39c6beafc2e28106b86e0edd
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57701
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Move the bulk of the constant evaulation logic out of transform::FoldConstants and into Resolver and sem::Expression.
transform::FoldConstants now replace TypeConstructor nodes that have a constant value on the expression.
This is ground work to:
* Cleaning up the HLSL uniform buffer indexing, which is `/` and `%` arithmatic heavy
* Prepares us to handle `constexpr` when it lands in the spec
* Provide a centralized place to do constant evaluation, instead of the
having similar logic scattered around the codebase.
Change-Id: I3e2f542be692046a8d243b62a82556db519953e7
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57426
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: James Price <jrprice@google.com>
This reverts commit fd5829e5ea.
Reason for revert: Temporarily reverting as this is preventing a tint->dawn roll, which is needed to fix the dawn->chrome roll.
Original change's description:
> Validate storage class constraints
>
> As defined by https://gpuweb.github.io/gpuweb/wgsl/#storage-class-layout-constraints
>
> Bug: tint:643
> Change-Id: I9c78ba69a792a80c263a17b0a6e9b4810fdb7f30
> Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/56780
> Kokoro: Kokoro <noreply+kokoro@google.com>
> Commit-Queue: Antonio Maiorano <amaiorano@google.com>
> Reviewed-by: Ben Clayton <bclayton@google.com>
TBR=bclayton@google.com,amaiorano@google.com,noreply+kokoro@google.com,tint-scoped@luci-project-accounts.iam.gserviceaccount.com
Change-Id: I8dbd0e46b3e3291ef08797a196d0d9abd2a78845
No-Presubmit: true
No-Tree-Checks: true
No-Try: true
Bug: tint:643
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57704
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>
... once the maximum number of errors have been reached.
https://dawn-review.googlesource.com/c/tint/+/56070 introduced maybe_set_synchronized(), which only set synchronized_ when the number of errors reported was less than max_errors_, but it seems the fuzzers have found ways to generate an excessive number of errors that keep the parser synchronized.
Revert 56070, and instead check the synchronized state along with the error count for every unbounded loop in the parser.
Fixed: chromium:1226655
Fixed: chromium:1226379
Change-Id: I178d758ac1424d4d19923fe6a3d9e123879b9eae
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57427
Auto-Submit: Ben Clayton <bclayton@google.com>
Commit-Queue: James Price <jrprice@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Report whether one was generated so that Dawn knows to use the
`-fpreserve-invariance` compiler option.
Bug: tint:772
Change-Id: Ife1eb05265646727dc864f12f983781af4df3777
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57644
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Make sure the other backends ICE on unrecognized attributes.
Add E2E tests, currently skipped for the other backends.
Bug: tint:772
Change-Id: I4e68d111ff79b19ebb6c574058a91debcb746011
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57642
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This is only valid on structure members and entry point return
types. Additionally, this can only be applied to a position builtin.
Bug: tint:772
Change-Id: Iffd774ca768ab66373678ecf0eb57c766767e92b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57641
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
This CL removes the old token `Is<Name>()` methods which are no longer
being used in the parser or lexer. (Some of them were used in tests but
the generic `Is(<type>)` is used instead.
New `peek` methods are added which will peek if the given token comes
either next or at the given index.
Change-Id: I8d15f2a42889f981d273b6459d20c4202db1ae32
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57540
Auto-Submit: dan sinclair <dan.sinclair@gmail.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
By generating a helper function for these, we can keep the atomic expression pre-statement-free. This can help prevent for-loops from being transformed into while loops.
Change-Id: Id034ea5ea9be601661ddb78db973015d845c420f
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57463
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
By generating a helper function for these, we can keep the atomic expression pre-statement-free. This can help prevent for-loops from being transformed into while loops, which can upset FXC.
We can't do the same for workgroup storage atomics, as the InterlockedXXX() methods have the workgroup-storage expression as the first argument, and I'm not aware of any way to make a user-declared parameter be `groupshared`.
Change-Id: I8669127a58dc9cae95ce316523029064b5c9b5fa
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57462
Commit-Queue: James Price <jrprice@google.com>
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
And remove the u32 overload of frexp (it's not in the spec).
Brings the number of failing tint end to end tests for MSL down to 19/1098.
The WG still haven't found consensus on reworking these two intrinsics.
It's very likely that their signature will change so that they return a structure instead of returning a value and outputing another as a pointer.
Until the WG makes a decision, let's implement these according to the current spec.
Some overloads are still failing due to MSL missing overloads of the pointer parameter being in the `threadgroup` address space.
I'm holding off fixing these until we know what's happening with these intrinsics.
See also:
https://github.com/gpuweb/gpuweb/issues/1480https://github.com/gpuweb/gpuweb/issues/1846
Change-Id: Ib6764e6659d840db41bc65fed2b8b283d1056c3d
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/57421
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>
Reviewed-by: James Price <jrprice@google.com>
Removes the unneeded texture_external overload of textureSample from
intrinsics.def.
Bug: tint:858
Change-Id: I49935bae47542af7b440c74c96418e005daf9e19
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/53940
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Ben Clayton <bclayton@google.com>