mirror of
https://github.com/encounter/dawn-cmake.git
synced 2025-07-08 14:15:58 +00:00
1ca6fbad8f
MSL has a limit on the number of threadgroup memory arguments, so use a struct to support an arbitrary number of workgroup variables. This commit introduces a `State` object to this transform, which is used to track which structs have been cloned eagerly, in order to avoid duplicating them. Bug: tint:938 Change-Id: Ia467db186e176a08f160455eab5fd3b3662f56b8 Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/65360 Auto-Submit: James Price <jrprice@google.com> Kokoro: James Price <jrprice@google.com> Commit-Queue: James Price <jrprice@google.com> Reviewed-by: Ben Clayton <bclayton@google.com>
21 lines
541 B
Plaintext
21 lines
541 B
Plaintext
#include <metal_stdlib>
|
|
|
|
using namespace metal;
|
|
struct tint_symbol_4 {
|
|
float2x3 v;
|
|
};
|
|
|
|
void tint_symbol_inner(uint local_invocation_index, threadgroup float2x3* const tint_symbol_1) {
|
|
{
|
|
*(tint_symbol_1) = float2x3();
|
|
}
|
|
threadgroup_barrier(mem_flags::mem_threadgroup);
|
|
(void) *(tint_symbol_1);
|
|
}
|
|
|
|
kernel void tint_symbol(threadgroup tint_symbol_4* tint_symbol_3 [[threadgroup(0)]], uint local_invocation_index [[thread_index_in_threadgroup]]) {
|
|
tint_symbol_inner(local_invocation_index, &((*(tint_symbol_3)).v));
|
|
return;
|
|
}
|
|
|