encounter/dawn-cmake
1
0
Fork 0
mirror of https://github.com/encounter/dawn-cmake.git synced 2025-12-16 00:17:03 +00:00
Code Issues Packages Projects Releases Wiki Activity

ast: Keep style consistent

Browse Source 
Methods and functions are `CamelCase()`
Public fields are `snake_case` with no trailing `_`
Private fields are `snake_case` with a trailing `_`

Remove pointless getters on fully immutable fields.
They provide no value, and just add `()` noise on use.

Remove unused methods.

Bug: tint:1231
Change-Id: If32efd039df48938efd5bc2186d51fe4853e9840
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/66600
Reviewed-by: David Neto <dneto@google.com>
Commit-Queue: Ben Clayton <bclayton@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
This commit is contained in:
Ben Clayton
2021-10-15 17:33:10 +00:00
committed by Tint LUCI CQ
parent b3e6c0d62c
commit 4f3ff57c28
337 changed files with 4287 additions and 4803 deletions
Download Patch File Download Diff File Expand all files Collapse all files

10
src/transform/array_length_from_uniform.cc
View File

@@ -106,21 +106,21 @@ void ArrayLengthFromUniform::Run(CloneContext& ctx,
// We assume that the argument to `arrayLength` has the form
// `&resource.array`, which requires that `InlinePointerLets` and
// `Simplify` have been run before this transform.
auto* param = call_expr->args()[0]->As<ast::UnaryOpExpression>();
if (!param || param->op() != ast::UnaryOp::kAddressOf) {
auto* param = call_expr->args[0]->As<ast::UnaryOpExpression>();
if (!param || param->op != ast::UnaryOp::kAddressOf) {
TINT_ICE(Transform, ctx.dst->Diagnostics())
<< "expected form of arrayLength argument to be "
"&resource.array";
break;
}
auto* accessor = param->expr()->As<ast::MemberAccessorExpression>();
auto* accessor = param->expr->As<ast::MemberAccessorExpression>();
if (!accessor) {
TINT_ICE(Transform, ctx.dst->Diagnostics())
<< "expected form of arrayLength argument to be "
"&resource.array";
break;
}
auto* storage_buffer_expr = accessor->structure();
auto* storage_buffer_expr = accessor->structure;
auto* storage_buffer_sem =
sem.Get(storage_buffer_expr)->As<sem::VariableUser>();
if (!storage_buffer_sem) {
@@ -151,7 +151,7 @@ void ArrayLengthFromUniform::Run(CloneContext& ctx,
// Load the total storage buffer size from the UBO.
uint32_t array_index = idx_itr->second / 4;
auto* vec_expr = ctx.dst->IndexAccessor(
ctx.dst->MemberAccessor(get_ubo()->symbol(), kBufferSizeMemberName),
ctx.dst->MemberAccessor(get_ubo()->symbol, kBufferSizeMemberName),
array_index);
uint32_t vec_index = idx_itr->second % 4;
auto* total_storage_buffer_size =

24
src/transform/binding_remapper.cc
View File

@@ -65,9 +65,9 @@ void BindingRemapper::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
auto* func = ctx.src->Sem().Get(func_ast);
std::unordered_map<sem::BindingPoint, int> binding_point_counts;
for (auto* var : func->ReferencedModuleVariables()) {
if (auto binding_point = var->Declaration()->binding_point()) {
BindingPoint from{binding_point.group->value(),
binding_point.binding->value()};
if (auto binding_point = var->Declaration()->BindingPoint()) {
BindingPoint from{binding_point.group->value,
binding_point.binding->value};
auto bp_it = remappings->binding_points.find(from);
if (bp_it != remappings->binding_points.end()) {
// Remapped
@@ -87,17 +87,17 @@ void BindingRemapper::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
}
for (auto* var : ctx.src->AST().GlobalVariables()) {
if (auto binding_point = var->binding_point()) {
if (auto binding_point = var->BindingPoint()) {
// The original binding point
BindingPoint from{binding_point.group->value(),
binding_point.binding->value()};
BindingPoint from{binding_point.group->value,
binding_point.binding->value};
// The binding point after remapping
BindingPoint bp = from;
// Replace any group or binding decorations.
// Note: This has to be performed *before* remapping access controls, as
// `ctx.Clone(var->decorations())` depend on these replacements.
// `ctx.Clone(var->decorations)` depend on these replacements.
auto bp_it = remappings->binding_points.find(from);
if (bp_it != remappings->binding_points.end()) {
BindingPoint to = bp_it->second;
@@ -125,15 +125,15 @@ void BindingRemapper::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
ctx.dst->Diagnostics().add_error(
diag::System::Transform,
"cannot apply access control to variable with storage class " +
std::string(ast::str(sem->StorageClass())));
std::string(ast::ToString(sem->StorageClass())));
return;
}
auto* ty = sem->Type()->UnwrapRef();
ast::Type* inner_ty = CreateASTTypeFor(ctx, ty);
auto* new_var = ctx.dst->create<ast::Variable>(
ctx.Clone(var->source()), ctx.Clone(var->symbol()),
var->declared_storage_class(), ac, inner_ty, var->is_const(),
ctx.Clone(var->constructor()), ctx.Clone(var->decorations()));
ctx.Clone(var->source), ctx.Clone(var->symbol),
var->declared_storage_class, ac, inner_ty, var->is_const,
ctx.Clone(var->constructor), ctx.Clone(var->decorations));
ctx.Replace(var, new_var);
}
@@ -142,7 +142,7 @@ void BindingRemapper::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
auto* decoration =
ctx.dst->ASTNodes().Create<ast::DisableValidationDecoration>(
ctx.dst->ID(), ast::DisabledValidation::kBindingPointCollision);
ctx.InsertBefore(var->decorations(), *var->decorations().begin(),
ctx.InsertBefore(var->decorations, *var->decorations.begin(),
decoration);
}
}

25
src/transform/calculate_array_length.cc
View File

@@ -54,9 +54,8 @@ struct ArrayUsage {
} // namespace
CalculateArrayLength::BufferSizeIntrinsic::BufferSizeIntrinsic(
ProgramID program_id)
: Base(program_id) {}
CalculateArrayLength::BufferSizeIntrinsic::BufferSizeIntrinsic(ProgramID pid)
: Base(pid) {}
CalculateArrayLength::BufferSizeIntrinsic::~BufferSizeIntrinsic() = default;
std::string CalculateArrayLength::BufferSizeIntrinsic::InternalName() const {
return "intrinsic_buffer_size";
@@ -82,7 +81,7 @@ void CalculateArrayLength::Run(CloneContext& ctx, const DataMap&, DataMap&) {
return utils::GetOrCreate(buffer_size_intrinsics, buffer_type, [&] {
auto name = ctx.dst->Sym();
auto* buffer_typename =
ctx.dst->ty.type_name(ctx.Clone(buffer_type->Declaration()->name()));
ctx.dst->ty.type_name(ctx.Clone(buffer_type->Declaration()->name));
auto* disable_validation =
ctx.dst->ASTNodes().Create<ast::DisableValidationDecoration>(
ctx.dst->ID(),
@@ -134,14 +133,14 @@ void CalculateArrayLength::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// We can assume that the arrayLength() call has a single argument of
// the form: arrayLength(&X.Y) where X is an expression that resolves
// to the storage buffer structure, and Y is the runtime sized array.
auto* arg = call_expr->args()[0];
auto* arg = call_expr->args[0];
auto* address_of = arg->As<ast::UnaryOpExpression>();
if (!address_of || address_of->op() != ast::UnaryOp::kAddressOf) {
if (!address_of || address_of->op != ast::UnaryOp::kAddressOf) {
TINT_ICE(Transform, ctx.dst->Diagnostics())
<< "arrayLength() expected pointer to member access, got "
<< address_of->TypeInfo().name;
}
auto* array_expr = address_of->expr();
auto* array_expr = address_of->expr;
auto* accessor = array_expr->As<ast::MemberAccessorExpression>();
if (!accessor) {
@@ -151,7 +150,7 @@ void CalculateArrayLength::Run(CloneContext& ctx, const DataMap&, DataMap&) {
<< array_expr->TypeInfo().name;
break;
}
auto* storage_buffer_expr = accessor->structure();
auto* storage_buffer_expr = accessor->structure;
auto* storage_buffer_sem = sem.Get(storage_buffer_expr);
auto* storage_buffer_type =
storage_buffer_sem->Type()->UnwrapRef()->As<sem::Struct>();
@@ -201,7 +200,7 @@ void CalculateArrayLength::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// X.GetDimensions(ARGS..) by the writer
buffer_size, ctx.Clone(storage_buffer_expr),
ctx.dst->AddressOf(ctx.dst->Expr(
buffer_size_result->variable()->symbol()))));
buffer_size_result->variable->symbol))));
// Calculate actual array length
// total_storage_buffer_size - array_offset
@@ -213,16 +212,16 @@ void CalculateArrayLength::Run(CloneContext& ctx, const DataMap&, DataMap&) {
auto* array_length_var = ctx.dst->Decl(ctx.dst->Const(
name, ctx.dst->ty.u32(),
ctx.dst->Div(
ctx.dst->Sub(buffer_size_result->variable()->symbol(),
ctx.dst->Sub(buffer_size_result->variable->symbol,
array_offset),
array_stride)));
// Insert the array length calculations at the top of the block
ctx.InsertBefore(block->statements(), *block->begin(),
ctx.InsertBefore(block->statements, block->statements[0],
buffer_size_result);
ctx.InsertBefore(block->statements(), *block->begin(),
ctx.InsertBefore(block->statements, block->statements[0],
call_get_dims);
ctx.InsertBefore(block->statements(), *block->begin(),
ctx.InsertBefore(block->statements, block->statements[0],
array_length_var);
return name;
});

64
src/transform/canonicalize_entry_point_io.cc
View File

@@ -40,24 +40,24 @@ namespace {
// those with builtin attributes.
bool StructMemberComparator(const ast::StructMember* a,
const ast::StructMember* b) {
auto* a_loc = ast::GetDecoration<ast::LocationDecoration>(a->decorations());
auto* b_loc = ast::GetDecoration<ast::LocationDecoration>(b->decorations());
auto* a_blt = ast::GetDecoration<ast::BuiltinDecoration>(a->decorations());
auto* b_blt = ast::GetDecoration<ast::BuiltinDecoration>(b->decorations());
auto* a_loc = ast::GetDecoration<ast::LocationDecoration>(a->decorations);
auto* b_loc = ast::GetDecoration<ast::LocationDecoration>(b->decorations);
auto* a_blt = ast::GetDecoration<ast::BuiltinDecoration>(a->decorations);
auto* b_blt = ast::GetDecoration<ast::BuiltinDecoration>(b->decorations);
if (a_loc) {
if (!b_loc) {
// `a` has location attribute and `b` does not: `a` goes first.
return true;
}
// Both have location attributes: smallest goes first.
return a_loc->value() < b_loc->value();
return a_loc->value < b_loc->value;
} else {
if (b_loc) {
// `b` has location attribute and `a` does not: `b` goes first.
return false;
}
// Both are builtins: order doesn't matter, just use enum value.
return a_blt->value() < b_blt->value();
return a_blt->builtin < b_blt->builtin;
}
}
@@ -70,7 +70,7 @@ bool IsShaderIODecoration(const ast::Decoration* deco) {
// Returns true if `decos` contains a `sample_mask` builtin.
bool HasSampleMask(const ast::DecorationList& decos) {
auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(decos);
return builtin && builtin->value() == ast::Builtin::kSampleMask;
return builtin && builtin->builtin == ast::Builtin::kSampleMask;
}
} // namespace
@@ -163,7 +163,7 @@ struct CanonicalizeEntryPointIO::State {
// always decorated with `Flat`.
if (type->is_integer_scalar_or_vector() &&
ast::HasDecoration<ast::LocationDecoration>(attributes) &&
func_ast->pipeline_stage() == ast::PipelineStage::kFragment) {
func_ast->PipelineStage() == ast::PipelineStage::kFragment) {
attributes.push_back(ctx.dst->Interpolate(
ast::InterpolationType::kFlat, ast::InterpolationSampling::kNone));
}
@@ -220,7 +220,7 @@ struct CanonicalizeEntryPointIO::State {
if (cfg.shader_style == ShaderStyle::kSpirv &&
type->is_integer_scalar_or_vector() &&
ast::HasDecoration<ast::LocationDecoration>(attributes) &&
func_ast->pipeline_stage() == ast::PipelineStage::kVertex) {
func_ast->PipelineStage() == ast::PipelineStage::kVertex) {
attributes.push_back(ctx.dst->Interpolate(
ast::InterpolationType::kFlat, ast::InterpolationSampling::kNone));
}
@@ -242,14 +242,14 @@ struct CanonicalizeEntryPointIO::State {
// Remove the shader IO attributes from the inner function parameter, and
// attach them to the new object instead.
ast::DecorationList attributes;
for (auto* deco : param->Declaration()->decorations()) {
for (auto* deco : param->Declaration()->decorations) {
if (IsShaderIODecoration(deco)) {
ctx.Remove(param->Declaration()->decorations(), deco);
ctx.Remove(param->Declaration()->decorations, deco);
attributes.push_back(ctx.Clone(deco));
}
}
auto name = ctx.src->Symbols().NameFor(param->Declaration()->symbol());
auto name = ctx.src->Symbols().NameFor(param->Declaration()->symbol);
auto* input_expr = AddInput(name, param->Type(), std::move(attributes));
inner_call_parameters.push_back(input_expr);
}
@@ -272,15 +272,15 @@ struct CanonicalizeEntryPointIO::State {
}
auto* member_ast = member->Declaration();
auto name = ctx.src->Symbols().NameFor(member_ast->symbol());
auto attributes = CloneShaderIOAttributes(member_ast->decorations());
auto name = ctx.src->Symbols().NameFor(member_ast->symbol);
auto attributes = CloneShaderIOAttributes(member_ast->decorations);
auto* input_expr = AddInput(name, member->Type(), std::move(attributes));
inner_struct_values.push_back(input_expr);
}
// Construct the original structure using the new shader input objects.
inner_call_parameters.push_back(ctx.dst->Construct(
ctx.Clone(param->Declaration()->type()), inner_struct_values));
ctx.Clone(param->Declaration()->type), inner_struct_values));
}
/// Process the entry point return type.
@@ -298,8 +298,8 @@ struct CanonicalizeEntryPointIO::State {
}
auto* member_ast = member->Declaration();
auto name = ctx.src->Symbols().NameFor(member_ast->symbol());
auto attributes = CloneShaderIOAttributes(member_ast->decorations());
auto name = ctx.src->Symbols().NameFor(member_ast->symbol);
auto attributes = CloneShaderIOAttributes(member_ast->decorations);
// Extract the original structure member.
AddOutput(name, member->Type(), std::move(attributes),
@@ -307,7 +307,7 @@ struct CanonicalizeEntryPointIO::State {
}
} else if (!inner_ret_type->Is<sem::Void>()) {
auto attributes =
CloneShaderIOAttributes(func_ast->return_type_decorations());
CloneShaderIOAttributes(func_ast->return_type_decorations);
// Propagate the non-struct return value as is.
AddOutput("value", func_sem->ReturnType(), std::move(attributes),
@@ -396,7 +396,7 @@ struct CanonicalizeEntryPointIO::State {
// Create the output struct object, assign its members, and return it.
auto* result_object =
ctx.dst->Var(wrapper_result, ctx.dst->ty.type_name(out_struct->name()));
ctx.dst->Var(wrapper_result, ctx.dst->ty.type_name(out_struct->name));
wrapper_body.push_back(ctx.dst->Decl(result_object));
wrapper_body.insert(wrapper_body.end(), assignments.begin(),
assignments.end());
@@ -435,31 +435,31 @@ struct CanonicalizeEntryPointIO::State {
ast::CallExpression* CallInnerFunction() {
// Add a suffix to the function name, as the wrapper function will take the
// original entry point name.
auto ep_name = ctx.src->Symbols().NameFor(func_ast->symbol());
auto ep_name = ctx.src->Symbols().NameFor(func_ast->symbol);
auto inner_name = ctx.dst->Symbols().New(ep_name + "_inner");
// Clone everything, dropping the function and return type attributes.
// The parameter attributes will have already been stripped during
// processing.
auto* inner_function = ctx.dst->create<ast::Function>(
inner_name, ctx.Clone(func_ast->params()),
ctx.Clone(func_ast->return_type()), ctx.Clone(func_ast->body()),
inner_name, ctx.Clone(func_ast->params),
ctx.Clone(func_ast->return_type), ctx.Clone(func_ast->body),
ast::DecorationList{}, ast::DecorationList{});
ctx.Replace(func_ast, inner_function);
// Call the function.
return ctx.dst->Call(inner_function->symbol(), inner_call_parameters);
return ctx.dst->Call(inner_function->symbol, inner_call_parameters);
}
/// Process the entry point function.
void Process() {
bool needs_fixed_sample_mask = false;
bool needs_vertex_point_size = false;
if (func_ast->pipeline_stage() == ast::PipelineStage::kFragment &&
if (func_ast->PipelineStage() == ast::PipelineStage::kFragment &&
cfg.fixed_sample_mask != 0xFFFFFFFF) {
needs_fixed_sample_mask = true;
}
if (func_ast->pipeline_stage() == ast::PipelineStage::kVertex &&
if (func_ast->PipelineStage() == ast::PipelineStage::kVertex &&
cfg.emit_vertex_point_size) {
needs_vertex_point_size = true;
}
@@ -506,7 +506,7 @@ struct CanonicalizeEntryPointIO::State {
// Process the original return type to determine the outputs that the
// outer function needs to produce.
ProcessReturnType(func_sem->ReturnType(), inner_result->symbol());
ProcessReturnType(func_sem->ReturnType(), inner_result->symbol);
}
// Add a fixed sample mask, if necessary.
@@ -526,17 +526,17 @@ struct CanonicalizeEntryPointIO::State {
} else {
auto* output_struct = CreateOutputStruct();
wrapper_ret_type = [&, output_struct] {
return ctx.dst->ty.type_name(output_struct->name());
return ctx.dst->ty.type_name(output_struct->name);
};
}
}
// Create the wrapper entry point function.
// Take the name of the original entry point function.
auto name = ctx.Clone(func_ast->symbol());
auto name = ctx.Clone(func_ast->symbol);
auto* wrapper_func = ctx.dst->create<ast::Function>(
name, wrapper_ep_parameters, wrapper_ret_type(),
ctx.dst->Block(wrapper_body), ctx.Clone(func_ast->decorations()),
ctx.dst->Block(wrapper_body), ctx.Clone(func_ast->decorations),
ast::DecorationList{});
ctx.InsertAfter(ctx.src->AST().GlobalDeclarations(), func_ast,
wrapper_func);
@@ -558,10 +558,10 @@ void CanonicalizeEntryPointIO::Run(CloneContext& ctx,
// New structures will be created for each entry point, as necessary.
for (auto* ty : ctx.src->AST().TypeDecls()) {
if (auto* struct_ty = ty->As<ast::Struct>()) {
for (auto* member : struct_ty->members()) {
for (auto* deco : member->decorations()) {
for (auto* member : struct_ty->members) {
for (auto* deco : member->decorations) {
if (IsShaderIODecoration(deco)) {
ctx.Remove(member->decorations(), deco);
ctx.Remove(member->decorations, deco);
}
}
}

54
src/transform/decompose_memory_access.cc
View File

@@ -332,11 +332,11 @@ struct DecomposeMemoryAccess::State {
/// @returns an Offset for the given ast::Expression
const Offset* ToOffset(ast::Expression* expr) {
if (auto* scalar = expr->As<ast::ScalarConstructorExpression>()) {
if (auto* u32 = scalar->literal()->As<ast::UintLiteral>()) {
return offsets_.Create<OffsetLiteral>(u32->value());
} else if (auto* i32 = scalar->literal()->As<ast::SintLiteral>()) {
if (i32->value() > 0) {
return offsets_.Create<OffsetLiteral>(i32->value());
if (auto* u32 = scalar->literal->As<ast::UintLiteral>()) {
return offsets_.Create<OffsetLiteral>(u32->value);
} else if (auto* i32 = scalar->literal->As<ast::SintLiteral>()) {
if (i32->value > 0) {
return offsets_.Create<OffsetLiteral>(i32->value);
}
}
}
@@ -625,7 +625,7 @@ struct DecomposeMemoryAccess::State {
for (auto* member : str->Members()) {
auto* offset = b.Add("offset", member->Offset());
auto* access = b.MemberAccessor(
"value", ctx.Clone(member->Declaration()->symbol()));
"value", ctx.Clone(member->Declaration()->symbol));
Symbol store =
StoreFunc(buf_ty, member->Type()->UnwrapRef(), var_user);
auto* call = b.Call(store, "buffer", offset, access);
@@ -697,16 +697,16 @@ struct DecomposeMemoryAccess::State {
ast::DecorationList{});
b.AST().AddFunction(func);
return func->symbol();
return func->symbol;
});
}
};
DecomposeMemoryAccess::Intrinsic::Intrinsic(ProgramID program_id,
DecomposeMemoryAccess::Intrinsic::Intrinsic(ProgramID pid,
Op o,
ast::StorageClass sc,
DataType ty)
: Base(program_id), op(o), storage_class(sc), type(ty) {}
: Base(pid), op(o), storage_class(sc), type(ty) {}
DecomposeMemoryAccess::Intrinsic::~Intrinsic() = default;
std::string DecomposeMemoryAccess::Intrinsic::InternalName() const {
std::stringstream ss;
@@ -837,7 +837,7 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
auto* accessor_sem = sem.Get(accessor);
if (auto* swizzle = accessor_sem->As<sem::Swizzle>()) {
if (swizzle->Indices().size() == 1) {
if (auto access = state.TakeAccess(accessor->structure())) {
if (auto access = state.TakeAccess(accessor->structure)) {
auto* vec_ty = access.type->As<sem::Vector>();
auto* offset =
state.Mul(vec_ty->type()->Size(), swizzle->Indices()[0]);
@@ -849,9 +849,9 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
}
}
} else {
if (auto access = state.TakeAccess(accessor->structure())) {
if (auto access = state.TakeAccess(accessor->structure)) {
auto* str_ty = access.type->As<sem::Struct>();
auto* member = str_ty->FindMember(accessor->member()->symbol());
auto* member = str_ty->FindMember(accessor->member->symbol);
auto offset = member->Offset();
state.AddAccess(accessor, {
access.var,
@@ -864,10 +864,10 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
}
if (auto* accessor = node->As<ast::ArrayAccessorExpression>()) {
if (auto access = state.TakeAccess(accessor->array())) {
if (auto access = state.TakeAccess(accessor->array)) {
// X[Y]
if (auto* arr = access.type->As<sem::Array>()) {
auto* offset = state.Mul(arr->Stride(), accessor->idx_expr());
auto* offset = state.Mul(arr->Stride(), accessor->index);
state.AddAccess(accessor, {
access.var,
state.Add(access.offset, offset),
@@ -876,8 +876,7 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
continue;
}
if (auto* vec_ty = access.type->As<sem::Vector>()) {
auto* offset =
state.Mul(vec_ty->type()->Size(), accessor->idx_expr());
auto* offset = state.Mul(vec_ty->type()->Size(), accessor->index);
state.AddAccess(accessor, {
access.var,
state.Add(access.offset, offset),
@@ -886,8 +885,7 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
continue;
}
if (auto* mat_ty = access.type->As<sem::Matrix>()) {
auto* offset =
state.Mul(mat_ty->ColumnStride(), accessor->idx_expr());
auto* offset = state.Mul(mat_ty->ColumnStride(), accessor->index);
state.AddAccess(accessor, {
access.var,
state.Add(access.offset, offset),
@@ -899,9 +897,9 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
}
if (auto* op = node->As<ast::UnaryOpExpression>()) {
if (op->op() == ast::UnaryOp::kAddressOf) {
if (op->op == ast::UnaryOp::kAddressOf) {
// &X
if (auto access = state.TakeAccess(op->expr())) {
if (auto access = state.TakeAccess(op->expr)) {
// HLSL does not support pointers, so just take the access from the
// reference and place it on the pointer.
state.AddAccess(op, access);
@@ -913,7 +911,7 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
if (auto* assign = node->As<ast::AssignmentStatement>()) {
// X = Y
// Move the LHS access to a store.
if (auto lhs = state.TakeAccess(assign->lhs())) {
if (auto lhs = state.TakeAccess(assign->lhs)) {
state.stores.emplace_back(Store{assign, lhs});
}
}
@@ -927,8 +925,8 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// may refer to a structure holding a runtime array, which cannot be
// loaded. Instead replace X with the underlying storage / uniform
// buffer variable.
if (auto access = state.TakeAccess(call_expr->args()[0])) {
ctx.Replace(call_expr->args()[0], [=, &ctx] {
if (auto access = state.TakeAccess(call_expr->args[0])) {
ctx.Replace(call_expr->args[0], [=, &ctx] {
return ctx.CloneWithoutTransform(access.var->Declaration());
});
}
@@ -938,11 +936,11 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// arrayLength(X)
// Don't convert X into a load, this intrinsic actually requires the
// real pointer.
state.TakeAccess(call_expr->args()[0]);
state.TakeAccess(call_expr->args[0]);
continue;
}
if (intrinsic->IsAtomic()) {
if (auto access = state.TakeAccess(call_expr->args()[0])) {
if (auto access = state.TakeAccess(call_expr->args[0])) {
// atomic___(X)
ctx.Replace(call_expr, [=, &ctx, &state] {
auto* buf = access.var->Declaration();
@@ -954,8 +952,8 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
access.var->As<sem::VariableUser>());
ast::ExpressionList args{ctx.Clone(buf), offset};
for (size_t i = 1; i < call_expr->args().size(); i++) {
auto* arg = call_expr->args()[i];
for (size_t i = 1; i < call_expr->args.size(); i++) {
auto* arg = call_expr->args[i];
args.emplace_back(ctx.Clone(arg));
}
return ctx.dst->Call(func, args);
@@ -992,7 +990,7 @@ void DecomposeMemoryAccess::Run(CloneContext& ctx, const DataMap&, DataMap&) {
auto* offset = store.target.offset->Build(ctx);
auto* buf_ty = store.target.var->Type()->UnwrapRef();
auto* el_ty = store.target.type->UnwrapRef();
auto* value = store.assignment->rhs();
auto* value = store.assignment->rhs;
Symbol func = state.StoreFunc(buf_ty, el_ty,
store.target.var->As<sem::VariableUser>());
auto* call = ctx.dst->Call(func, ctx.CloneWithoutTransform(buf), offset,

17
src/transform/decompose_strided_matrix.cc
View File

@@ -86,11 +86,11 @@ void GatherCustomStrideMatrixMembers(const Program* program, F&& callback) {
continue;
}
auto* deco = ast::GetDecoration<ast::StrideDecoration>(
member->Declaration()->decorations());
member->Declaration()->decorations);
if (!deco) {
continue;
}
uint32_t stride = deco->stride();
uint32_t stride = deco->stride;
if (matrix->ColumnStride() == stride) {
continue;
}
@@ -147,11 +147,11 @@ void DecomposeStridedMatrix::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ctx.ReplaceAll(
[&](ast::ArrayAccessorExpression* expr) -> ast::ArrayAccessorExpression* {
if (auto* access =
ctx.src->Sem().Get<sem::StructMemberAccess>(expr->array())) {
ctx.src->Sem().Get<sem::StructMemberAccess>(expr->array)) {
auto it = decomposed.find(access->Member()->Declaration());
if (it != decomposed.end()) {
auto* obj = ctx.CloneWithoutTransform(expr->array());
auto* idx = ctx.Clone(expr->idx_expr());
auto* obj = ctx.CloneWithoutTransform(expr->array);
auto* idx = ctx.Clone(expr->index);
return ctx.dst->IndexAccessor(obj, idx);
}
}
@@ -165,8 +165,7 @@ void DecomposeStridedMatrix::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// ssbo.mat = mat_to_arr(m)
std::unordered_map<MatrixInfo, Symbol, MatrixInfo::Hasher> mat_to_arr;
ctx.ReplaceAll([&](ast::AssignmentStatement* stmt) -> ast::Statement* {
if (auto* access =
ctx.src->Sem().Get<sem::StructMemberAccess>(stmt->lhs())) {
if (auto* access = ctx.src->Sem().Get<sem::StructMemberAccess>(stmt->lhs)) {
auto it = decomposed.find(access->Member()->Declaration());
if (it == decomposed.end()) {
return nullptr;
@@ -196,8 +195,8 @@ void DecomposeStridedMatrix::Run(CloneContext& ctx, const DataMap&, DataMap&) {
});
return name;
});
auto* lhs = ctx.CloneWithoutTransform(stmt->lhs());
auto* rhs = ctx.dst->Call(fn, ctx.Clone(stmt->rhs()));
auto* lhs = ctx.CloneWithoutTransform(stmt->lhs);
auto* rhs = ctx.dst->Call(fn, ctx.Clone(stmt->rhs));
return ctx.dst->Assign(lhs, rhs);
}
return nullptr;

34
src/transform/external_texture_transform.cc
View File

@@ -53,43 +53,43 @@ void ExternalTextureTransform::Run(CloneContext& ctx,
// When a textureLoad or textureSampleLevel has been identified, check
// if the first parameter is an external texture.
if (auto* var =
sem.Get(call_expr->args()[0])->As<sem::VariableUser>()) {
sem.Get(call_expr->args[0])->As<sem::VariableUser>()) {
if (var->Variable()
->Type()
->UnwrapRef()
->Is<sem::ExternalTexture>()) {
if (intrinsic->Type() == sem::IntrinsicType::kTextureLoad &&
call_expr->args().size() != 2) {
call_expr->args.size() != 2) {
TINT_ICE(Transform, ctx.dst->Diagnostics())
<< "expected textureLoad call with a texture_external to "
"have 2 parameters, found "
<< call_expr->args().size() << " parameters";
<< call_expr->args.size() << " parameters";
}
if (intrinsic->Type() ==
sem::IntrinsicType::kTextureSampleLevel &&
call_expr->args().size() != 3) {
call_expr->args.size() != 3) {
TINT_ICE(Transform, ctx.dst->Diagnostics())
<< "expected textureSampleLevel call with a "
"texture_external to have 3 parameters, found "
<< call_expr->args().size() << " parameters";
<< call_expr->args.size() << " parameters";
}
// Replace the call with another that has the same parameters in
// addition to a level parameter (always zero for external
// textures).
auto* exp = ctx.Clone(call_expr->func());
auto* externalTextureParam = ctx.Clone(call_expr->args()[0]);
auto* exp = ctx.Clone(call_expr->func);
auto* externalTextureParam = ctx.Clone(call_expr->args[0]);
ast::ExpressionList params;
if (intrinsic->Type() == sem::IntrinsicType::kTextureLoad) {
auto* coordsParam = ctx.Clone(call_expr->args()[1]);
auto* coordsParam = ctx.Clone(call_expr->args[1]);
auto* levelParam = ctx.dst->Expr(0);
params = {externalTextureParam, coordsParam, levelParam};
} else if (intrinsic->Type() ==
sem::IntrinsicType::kTextureSampleLevel) {
auto* samplerParam = ctx.Clone(call_expr->args()[1]);
auto* coordsParam = ctx.Clone(call_expr->args()[2]);
auto* samplerParam = ctx.Clone(call_expr->args[1]);
auto* coordsParam = ctx.Clone(call_expr->args[2]);
auto* levelParam = ctx.dst->Expr(0.0f);
params = {externalTextureParam, samplerParam, coordsParam,
levelParam};
@@ -107,18 +107,18 @@ void ExternalTextureTransform::Run(CloneContext& ctx,
// Scan the AST nodes for external texture declarations.
for (auto* node : ctx.src->ASTNodes().Objects()) {
if (auto* var = node->As<ast::Variable>()) {
if (::tint::Is<ast::ExternalTexture>(var->type())) {
if (::tint::Is<ast::ExternalTexture>(var->type)) {
// Replace a single-plane external texture with a 2D, f32 sampled
// texture.
auto* newType = ctx.dst->ty.sampled_texture(ast::TextureDimension::k2d,
ctx.dst->ty.f32());
auto clonedSrc = ctx.Clone(var->source());
auto clonedSym = ctx.Clone(var->symbol());
auto* clonedConstructor = ctx.Clone(var->constructor());
auto clonedDecorations = ctx.Clone(var->decorations());
auto clonedSrc = ctx.Clone(var->source);
auto clonedSym = ctx.Clone(var->symbol);
auto* clonedConstructor = ctx.Clone(var->constructor);
auto clonedDecorations = ctx.Clone(var->decorations);
auto* newVar = ctx.dst->create<ast::Variable>(
clonedSrc, clonedSym, var->declared_storage_class(),
var->declared_access(), newType, var->is_const(), clonedConstructor,
clonedSrc, clonedSym, var->declared_storage_class,
var->declared_access, newType, var->is_const, clonedConstructor,
clonedDecorations);
ctx.Replace(var, newVar);

8
src/transform/first_index_offset.cc
View File

@@ -80,9 +80,9 @@ void FirstIndexOffset::Run(CloneContext& ctx,
// parameters) or structure member accesses.
for (auto* node : ctx.src->ASTNodes().Objects()) {
if (auto* var = node->As<ast::Variable>()) {
for (ast::Decoration* dec : var->decorations()) {
for (ast::Decoration* dec : var->decorations) {
if (auto* builtin_dec = dec->As<ast::BuiltinDecoration>()) {
ast::Builtin builtin = builtin_dec->value();
ast::Builtin builtin = builtin_dec->builtin;
if (builtin == ast::Builtin::kVertexIndex) {
auto* sem_var = ctx.src->Sem().Get(var);
builtin_vars.emplace(sem_var, kFirstVertexName);
@@ -97,9 +97,9 @@ void FirstIndexOffset::Run(CloneContext& ctx,
}
}
if (auto* member = node->As<ast::StructMember>()) {
for (ast::Decoration* dec : member->decorations()) {
for (ast::Decoration* dec : member->decorations) {
if (auto* builtin_dec = dec->As<ast::BuiltinDecoration>()) {
ast::Builtin builtin = builtin_dec->value();
ast::Builtin builtin = builtin_dec->builtin;
if (builtin == ast::Builtin::kVertexIndex) {
auto* sem_mem = ctx.src->Sem().Get(member);
builtin_members.emplace(sem_mem, kFirstVertexName);

4
src/transform/fold_constants.cc
View File

@@ -51,7 +51,7 @@ void FoldConstants::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// If original ctor expression had no init values, don't replace the
// expression
if (ctor->values().size() == 0) {
if (ctor->values.size() == 0) {
return nullptr;
}
@@ -68,7 +68,7 @@ void FoldConstants::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// create it with 3. So what we do is construct with vec_size args,
// except if the original vector was single-value initialized, in
// which case, we only construct with one arg again.
uint32_t ctor_size = (ctor->values().size() == 1) ? 1 : vec_size;
uint32_t ctor_size = (ctor->values.size() == 1) ? 1 : vec_size;
ast::ExpressionList ctors;
for (uint32_t i = 0; i < ctor_size; ++i) {

12
src/transform/fold_trivial_single_use_lets.cc
View File

@@ -32,11 +32,11 @@ ast::VariableDeclStatement* AsTrivialLetDecl(ast::Statement* stmt) {
if (!var_decl) {
return nullptr;
}
auto* var = var_decl->variable();
if (!var->is_const()) {
auto* var = var_decl->variable;
if (!var->is_const) {
return nullptr;
}
auto* ctor = var->constructor();
auto* ctor = var->constructor;
if (!IsAnyOf<ast::IdentifierExpression, ast::ScalarConstructorExpression>(
ctor)) {
return nullptr;
@@ -55,11 +55,11 @@ void FoldTrivialSingleUseLets::Run(CloneContext& ctx,
DataMap&) {
for (auto* node : ctx.src->ASTNodes().Objects()) {
if (auto* block = node->As<ast::BlockStatement>()) {
auto& stmts = block->statements();
auto& stmts = block->statements;
for (size_t stmt_idx = 0; stmt_idx < stmts.size(); stmt_idx++) {
auto* stmt = stmts[stmt_idx];
if (auto* let_decl = AsTrivialLetDecl(stmt)) {
auto* let = let_decl->variable();
auto* let = let_decl->variable;
auto* sem_let = ctx.src->Sem().Get(let);
auto& users = sem_let->Users();
if (users.size() != 1) {
@@ -73,7 +73,7 @@ void FoldTrivialSingleUseLets::Run(CloneContext& ctx,
if (user_stmt == stmts[i]) {
auto* user_expr = user->Declaration();
ctx.Remove(stmts, let_decl);
ctx.Replace(user_expr, ctx.Clone(let->constructor()));
ctx.Replace(user_expr, ctx.Clone(let->constructor));
}
if (!AsTrivialLetDecl(stmts[i])) {
// Stop if we hit a statement that isn't the single use of the

8
src/transform/for_loop_to_loop.cc
View File

@@ -28,7 +28,7 @@ ForLoopToLoop::~ForLoopToLoop() = default;
void ForLoopToLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ctx.ReplaceAll([&](ast::ForLoopStatement* for_loop) -> ast::Statement* {
ast::StatementList stmts;
if (auto* cond = for_loop->condition()) {
if (auto* cond = for_loop->condition) {
// !condition
auto* not_cond = ctx.dst->create<ast::UnaryOpExpression>(
ast::UnaryOp::kNot, ctx.Clone(cond));
@@ -39,19 +39,19 @@ void ForLoopToLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// if (!condition) { break; }
stmts.emplace_back(ctx.dst->If(not_cond, break_body));
}
for (auto* stmt : for_loop->body()->statements()) {
for (auto* stmt : for_loop->body->statements) {
stmts.emplace_back(ctx.Clone(stmt));
}
ast::BlockStatement* continuing = nullptr;
if (auto* cont = for_loop->continuing()) {
if (auto* cont = for_loop->continuing) {
continuing = ctx.dst->Block(ctx.Clone(cont));
}
auto* body = ctx.dst->Block(stmts);
auto* loop = ctx.dst->create<ast::LoopStatement>(body, continuing);
if (auto* init = for_loop->initializer()) {
if (auto* init = for_loop->initializer) {
return ctx.dst->Block(ctx.Clone(init), loop);
}

24
src/transform/inline_pointer_lets.cc
View File

@@ -44,21 +44,21 @@ void CollectSavedArrayIndices(const Program* program,
ast::Expression* expr,
F&& cb) {
if (auto* a = expr->As<ast::ArrayAccessorExpression>()) {
CollectSavedArrayIndices(program, a->array(), cb);
CollectSavedArrayIndices(program, a->array, cb);
if (!a->idx_expr()->Is<ast::ScalarConstructorExpression>()) {
cb(a->idx_expr());
if (!a->index->Is<ast::ScalarConstructorExpression>()) {
cb(a->index);
}
return;
}
if (auto* m = expr->As<ast::MemberAccessorExpression>()) {
CollectSavedArrayIndices(program, m->structure(), cb);
CollectSavedArrayIndices(program, m->structure, cb);
return;
}
if (auto* u = expr->As<ast::UnaryOpExpression>()) {
CollectSavedArrayIndices(program, u->expr(), cb);
CollectSavedArrayIndices(program, u->expr, cb);
return;
}
@@ -119,7 +119,7 @@ void InlinePointerLets::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// TINT_SCOPED_ASSIGNMENT provides a stack of PtrLet*, this is
// required to handle the 'chaining' of inlined `let`s.
TINT_SCOPED_ASSIGNMENT(current_ptr_let, ptr_let);
return ctx.Clone(var->constructor());
return ctx.Clone(var->constructor);
}
}
}
@@ -131,11 +131,11 @@ void InlinePointerLets::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// permitted.
for (auto* node : ctx.src->ASTNodes().Objects()) {
if (auto* let = node->As<ast::VariableDeclStatement>()) {
if (!let->variable()->is_const()) {
if (!let->variable->is_const) {
continue; // Not a `let` declaration. Ignore.
}
auto* var = ctx.src->Sem().Get(let->variable());
auto* var = ctx.src->Sem().Get(let->variable);
if (!var->Type()->Is<sem::Pointer>()) {
continue; // Not a pointer type. Ignore.
}
@@ -149,12 +149,12 @@ void InlinePointerLets::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// Scan the initializer expression for array index expressions that need
// to be hoist to temporary "saved" variables.
CollectSavedArrayIndices(
ctx.src, var->Declaration()->constructor(),
ctx.src, var->Declaration()->constructor,
[&](ast::Expression* idx_expr) {
// We have a sub-expression that needs to be saved.
// Create a new variable
auto saved_name = ctx.dst->Symbols().New(
ctx.src->Symbols().NameFor(var->Declaration()->symbol()) +
ctx.src->Symbols().NameFor(var->Declaration()->symbol) +
"_save");
auto* saved = ctx.dst->Decl(
ctx.dst->Const(saved_name, nullptr, ctx.Clone(idx_expr)));
@@ -165,7 +165,7 @@ void InlinePointerLets::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// Note that repeated calls to InsertAfter() with the same `after`
// argument will result in nodes to inserted in the order the calls
// are made (last call is inserted last).
ctx.InsertAfter(block->statements(), let, saved);
ctx.InsertAfter(block->statements, let, saved);
// Record the substitution of `idx_expr` to the saved variable with
// the symbol `saved_name`. This will be used by the ReplaceAll()
// handler above.
@@ -174,7 +174,7 @@ void InlinePointerLets::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// Record the pointer-typed `let` declaration.
// This will be used by the ReplaceAll() handler above.
ptr_lets.emplace(let->variable(), std::move(ptr_let));
ptr_lets.emplace(let->variable, std::move(ptr_let));
// As the original `let` declaration will be fully inlined, there's no
// need for the original declaration to exist. Remove it.
RemoveStatement(ctx, let);

34
src/transform/loop_to_for_loop.cc
View File

@@ -30,10 +30,10 @@ namespace transform {
namespace {
bool IsBlockWithSingleBreak(ast::BlockStatement* block) {
if (block->size() != 1) {
if (block->statements.size() != 1) {
return false;
}
return block->statements()[0]->Is<ast::BreakStatement>();
return block->statements[0]->Is<ast::BreakStatement>();
}
bool IsVarUsedByStmt(const sem::Info& sem,
@@ -67,7 +67,7 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// Examples:
// loop { if (condition) { break; } ... }
// loop { if (condition) {} else { break; } ... }
auto& stmts = loop->body()->statements();
auto& stmts = loop->body->statements;
if (stmts.empty()) {
return nullptr;
}
@@ -77,13 +77,12 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
}
bool negate_condition = false;
if (IsBlockWithSingleBreak(if_stmt->body()) &&
if_stmt->else_statements().empty()) {
if (IsBlockWithSingleBreak(if_stmt->body) &&
if_stmt->else_statements.empty()) {
negate_condition = true;
} else if (if_stmt->body()->empty() &&
if_stmt->else_statements().size() == 1 &&
if_stmt->else_statements()[0]->condition() == nullptr &&
IsBlockWithSingleBreak(if_stmt->else_statements()[0]->body())) {
} else if (if_stmt->body->Empty() && if_stmt->else_statements.size() == 1 &&
if_stmt->else_statements[0]->condition == nullptr &&
IsBlockWithSingleBreak(if_stmt->else_statements[0]->body)) {
negate_condition = false;
} else {
return nullptr;
@@ -92,12 +91,12 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// The continuing block must be empty or contain a single, assignment or
// function call statement.
ast::Statement* continuing = nullptr;
if (auto* loop_cont = loop->continuing()) {
if (loop_cont->statements().size() != 1) {
if (auto* loop_cont = loop->continuing) {
if (loop_cont->statements.size() != 1) {
return nullptr;
}
continuing = loop_cont->statements()[0];
continuing = loop_cont->statements[0];
if (!continuing
->IsAnyOf<ast::AssignmentStatement, ast::CallStatement>()) {
return nullptr;
@@ -105,10 +104,9 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
// And the continuing statement must not use any of the variables declared
// in the loop body.
for (auto* stmt : loop->body()->statements()) {
for (auto* stmt : loop->body->statements) {
if (auto* var_decl = stmt->As<ast::VariableDeclStatement>()) {
if (IsVarUsedByStmt(ctx.src->Sem(), var_decl->variable(),
continuing)) {
if (IsVarUsedByStmt(ctx.src->Sem(), var_decl->variable, continuing)) {
return nullptr;
}
}
@@ -117,7 +115,7 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
continuing = ctx.Clone(continuing);
}
auto* condition = ctx.Clone(if_stmt->condition());
auto* condition = ctx.Clone(if_stmt->condition);
if (negate_condition) {
condition = ctx.dst->create<ast::UnaryOpExpression>(ast::UnaryOp::kNot,
condition);
@@ -125,8 +123,8 @@ void LoopToForLoop::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ast::Statement* initializer = nullptr;
ctx.Remove(loop->body()->statements(), if_stmt);
auto* body = ctx.Clone(loop->body());
ctx.Remove(loop->body->statements, if_stmt);
auto* body = ctx.Clone(loop->body);
return ctx.dst->create<ast::ForLoopStatement>(initializer, condition,
continuing, body);
});

24
src/transform/module_scope_var_to_entry_point_param.cc
View File

@@ -127,10 +127,10 @@ struct ModuleScopeVarToEntryPointParam::State {
ident_to_address_of;
for (auto* node : ctx.src->ASTNodes().Objects()) {
auto* address_of = node->As<ast::UnaryOpExpression>();
if (!address_of || address_of->op() != ast::UnaryOp::kAddressOf) {
if (!address_of || address_of->op != ast::UnaryOp::kAddressOf) {
continue;
}
if (auto* ident = address_of->expr()->As<ast::IdentifierExpression>()) {
if (auto* ident = address_of->expr->As<ast::IdentifierExpression>()) {
ident_to_address_of[ident] = address_of;
}
}
@@ -180,10 +180,10 @@ struct ModuleScopeVarToEntryPointParam::State {
ctx.dst->ASTNodes().Create<ast::DisableValidationDecoration>(
ctx.dst->ID(),
ast::DisabledValidation::kEntryPointParameter);
auto decos = ctx.Clone(var->Declaration()->decorations());
auto decos = ctx.Clone(var->Declaration()->decorations);
decos.push_back(disable_validation);
auto* param = ctx.dst->Param(new_var_symbol, store_type(), decos);
ctx.InsertFront(func_ast->params(), param);
ctx.InsertFront(func_ast->params, param);
} else if (var->StorageClass() == ast::StorageClass::kWorkgroup &&
ContainsMatrix(var->Type())) {
// Due to a bug in the MSL compiler, we use a threadgroup memory
@@ -192,7 +192,7 @@ struct ModuleScopeVarToEntryPointParam::State {
// TODO(jrprice): Do this for all other workgroup variables too.
// Create a member in the workgroup parameter struct.
auto member = ctx.Clone(var->Declaration()->symbol());
auto member = ctx.Clone(var->Declaration()->symbol);
workgroup_parameter_members.push_back(
ctx.dst->Member(member, store_type()));
CloneStructTypes(var->Type()->UnwrapRef());
@@ -205,7 +205,7 @@ struct ModuleScopeVarToEntryPointParam::State {
ctx.dst->ty.pointer(
store_type(), ast::StorageClass::kWorkgroup),
member_ptr);
ctx.InsertFront(func_ast->body()->statements(),
ctx.InsertFront(func_ast->body->statements,
ctx.dst->Decl(local_var));
is_pointer = true;
} else {
@@ -216,11 +216,11 @@ struct ModuleScopeVarToEntryPointParam::State {
ctx.dst->ASTNodes().Create<ast::DisableValidationDecoration>(
ctx.dst->ID(),
ast::DisabledValidation::kIgnoreStorageClass);
auto* constructor = ctx.Clone(var->Declaration()->constructor());
auto* constructor = ctx.Clone(var->Declaration()->constructor);
auto* local_var = ctx.dst->Var(
new_var_symbol, store_type(), var->StorageClass(), constructor,
ast::DecorationList{disable_validation});
ctx.InsertFront(func_ast->body()->statements(),
ctx.InsertFront(func_ast->body->statements,
ctx.dst->Decl(local_var));
}
} else {
@@ -240,7 +240,7 @@ struct ModuleScopeVarToEntryPointParam::State {
ast::DisabledValidation::kIgnoreInvalidPointerArgument));
}
ctx.InsertBack(
func_ast->params(),
func_ast->params,
ctx.dst->Param(new_var_symbol, param_type, attributes));
}
@@ -282,12 +282,12 @@ struct ModuleScopeVarToEntryPointParam::State {
ctx.dst->ID(), ast::DisabledValidation::kEntryPointParameter);
auto* param =
ctx.dst->Param(workgroup_param(), param_type, {disable_validation});
ctx.InsertFront(func_ast->params(), param);
ctx.InsertFront(func_ast->params, param);
}
// Pass the variables as pointers to any functions that need them.
for (auto* call : calls_to_replace[func_ast]) {
auto* target = ctx.src->AST().Functions().Find(call->func()->symbol());
auto* target = ctx.src->AST().Functions().Find(call->func->symbol);
auto* target_sem = ctx.src->Sem().Get(target);
// Add new arguments for any variables that are needed by the callee.
@@ -305,7 +305,7 @@ struct ModuleScopeVarToEntryPointParam::State {
if (is_entry_point && !is_handle && !is_workgroup_matrix) {
arg = ctx.dst->AddressOf(arg);
}
ctx.InsertBack(call->args(), arg);
ctx.InsertBack(call->args, arg);
}
}
}

18
src/transform/num_workgroups_from_uniform.cc
View File

@@ -73,7 +73,7 @@ void NumWorkgroupsFromUniform::Run(CloneContext& ctx,
std::unordered_set<Accessor, Accessor::Hasher> to_replace;
for (auto* func : ctx.src->AST().Functions()) {
// num_workgroups is only valid for compute stages.
if (func->pipeline_stage() != ast::PipelineStage::kCompute) {
if (func->PipelineStage() != ast::PipelineStage::kCompute) {
continue;
}
@@ -87,8 +87,8 @@ void NumWorkgroupsFromUniform::Run(CloneContext& ctx,
for (auto* member : str->Members()) {
auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(
member->Declaration()->decorations());
if (!builtin || builtin->value() != ast::Builtin::kNumWorkgroups) {
member->Declaration()->decorations);
if (!builtin || builtin->builtin != ast::Builtin::kNumWorkgroups) {
continue;
}
@@ -96,18 +96,18 @@ void NumWorkgroupsFromUniform::Run(CloneContext& ctx,
// we will replace later. We currently have no way to get from the
// parameter directly to the member accessor expressions that use it.
to_replace.insert(
{param->Declaration()->symbol(), member->Declaration()->symbol()});
{param->Declaration()->symbol, member->Declaration()->symbol});
// Remove the struct member.
// The CanonicalizeEntryPointIO transform will have generated this
// struct uniquely for this particular entry point, so we know that
// there will be no other uses of this struct in the module and that we
// can safely modify it here.
ctx.Remove(str->Declaration()->members(), member->Declaration());
ctx.Remove(str->Declaration()->members, member->Declaration());
// If this is the only member, remove the struct and parameter too.
if (str->Members().size() == 1) {
ctx.Remove(func->params(), param->Declaration());
ctx.Remove(func->params, param->Declaration());
ctx.Remove(ctx.src->AST().GlobalDeclarations(), str->Declaration());
}
}
@@ -140,13 +140,13 @@ void NumWorkgroupsFromUniform::Run(CloneContext& ctx,
if (!accessor) {
continue;
}
auto* ident = accessor->structure()->As<ast::IdentifierExpression>();
auto* ident = accessor->structure->As<ast::IdentifierExpression>();
if (!ident) {
continue;
}
if (to_replace.count({ident->symbol(), accessor->member()->symbol()})) {
ctx.Replace(accessor, ctx.dst->MemberAccessor(get_ubo()->symbol(),
if (to_replace.count({ident->symbol, accessor->member->symbol})) {
ctx.Replace(accessor, ctx.dst->MemberAccessor(get_ubo()->symbol,
kNumWorkgroupsMemberName));
}
}

8
src/transform/pad_array_elements.cc
View File

@@ -118,7 +118,7 @@ void PadArrayElements::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ctx.ReplaceAll(
[&](ast::ArrayAccessorExpression* accessor) -> ast::Expression* {
if (auto* array = tint::As<sem::Array>(
sem.Get(accessor->array())->Type()->UnwrapRef())) {
sem.Get(accessor->array)->Type()->UnwrapRef())) {
if (pad(array)) {
// Array element is wrapped in a structure. Emit a member accessor
// to get to the actual array element.
@@ -136,11 +136,11 @@ void PadArrayElements::Run(CloneContext& ctx, const DataMap&, DataMap&) {
tint::As<sem::Array>(sem.Get(ctor)->Type()->UnwrapRef())) {
if (auto p = pad(array)) {
auto* arr_ty = p();
auto el_typename = arr_ty->type()->As<ast::TypeName>()->name();
auto el_typename = arr_ty->type->As<ast::TypeName>()->name;
ast::ExpressionList args;
args.reserve(ctor->values().size());
for (auto* arg : ctor->values()) {
args.reserve(ctor->values.size());
for (auto* arg : ctor->values) {
args.emplace_back(ctx.dst->Construct(
ctx.dst->create<ast::TypeName>(el_typename), ctx.Clone(arg)));
}

6
src/transform/promote_initializers_to_const_var.cc
View File

@@ -68,7 +68,7 @@ void PromoteInitializersToConstVar::Run(CloneContext& ctx,
auto* src_stmt = src_sem_stmt->Declaration();
if (auto* src_var_decl = src_stmt->As<ast::VariableDeclStatement>()) {
if (src_var_decl->variable()->constructor() == src_init) {
if (src_var_decl->variable->constructor == src_init) {
// This statement is just a variable declaration with the initializer
// as the constructor value. This is what we're attempting to
// transform to, and so ignore.
@@ -81,7 +81,7 @@ void PromoteInitializersToConstVar::Run(CloneContext& ctx,
// Create a new symbol for the constant
auto dst_symbol = ctx.dst->Sym();
// Clone the type
auto* dst_ty = ctx.Clone(src_init->type());
auto* dst_ty = ctx.Clone(src_init->type);
// Clone the initializer
auto* dst_init = ctx.Clone(src_init);
// Construct the constant that holds the hoisted initializer
@@ -92,7 +92,7 @@ void PromoteInitializersToConstVar::Run(CloneContext& ctx,
auto* dst_ident = ctx.dst->Expr(dst_symbol);
// Insert the constant before the usage
ctx.InsertBefore(src_sem_stmt->Block()->Declaration()->statements(),
ctx.InsertBefore(src_sem_stmt->Block()->Declaration()->statements,
src_stmt, dst_var_decl);
// Replace the inlined initializer with a reference to the constant
ctx.Replace(src_init, dst_ident);

12
src/transform/renamer.cc
View File

@@ -1115,11 +1115,11 @@ Output Renamer::Run(const Program* in, const DataMap& inputs) {
continue;
}
if (sem->Is<sem::Swizzle>()) {
preserve.emplace(member->member());
} else if (auto* str_expr = in->Sem().Get(member->structure())) {
preserve.emplace(member->member);
} else if (auto* str_expr = in->Sem().Get(member->structure)) {
if (auto* ty = str_expr->Type()->UnwrapRef()->As<sem::Struct>()) {
if (ty->Declaration() == nullptr) { // Builtin structure
preserve.emplace(member->member());
preserve.emplace(member->member);
}
}
}
@@ -1131,7 +1131,7 @@ Output Renamer::Run(const Program* in, const DataMap& inputs) {
continue;
}
if (sem->Target()->Is<sem::Intrinsic>()) {
preserve.emplace(call->func());
preserve.emplace(call->func);
}
}
}
@@ -1190,11 +1190,11 @@ Output Renamer::Run(const Program* in, const DataMap& inputs) {
ctx.ReplaceAll(
[&](ast::IdentifierExpression* ident) -> ast::IdentifierExpression* {
if (preserve.count(ident)) {
auto sym_in = ident->symbol();
auto sym_in = ident->symbol;
auto str = in->Symbols().NameFor(sym_in);
auto sym_out = out.Symbols().Register(str);
return ctx.dst->create<ast::IdentifierExpression>(
ctx.Clone(ident->source()), sym_out);
ctx.Clone(ident->source), sym_out);
}
return nullptr; // Clone ident. Uses the symbol remapping above.
});

26
src/transform/robustness.cc
View File

@@ -46,7 +46,7 @@ struct Robustness::State {
/// @return the clamped replacement expression, or nullptr if `expr` should be
/// cloned without changes.
ast::ArrayAccessorExpression* Transform(ast::ArrayAccessorExpression* expr) {
auto* ret_type = ctx.src->Sem().Get(expr->array())->Type()->UnwrapRef();
auto* ret_type = ctx.src->Sem().Get(expr->array)->Type()->UnwrapRef();
ProgramBuilder& b = *ctx.dst;
using u32 = ProgramBuilder::u32;
@@ -78,11 +78,11 @@ struct Robustness::State {
if (size.u32 == 0) {
if (!ret_type->Is<sem::Array>()) {
b.Diagnostics().add_error(diag::System::Transform,
"invalid 0 sized non-array", expr->source());
"invalid 0 sized non-array", expr->source);
return nullptr;
}
// Runtime sized array
auto* arr = ctx.Clone(expr->array());
auto* arr = ctx.Clone(expr->array);
size.expr = b.Call("arrayLength", b.AddressOf(arr));
}
@@ -101,7 +101,7 @@ struct Robustness::State {
Value idx; // index value
auto* idx_sem = ctx.src->Sem().Get(expr->idx_expr());
auto* idx_sem = ctx.src->Sem().Get(expr->index);
auto* idx_ty = idx_sem->Type()->UnwrapRef();
if (!idx_ty->IsAnyOf<sem::I32, sem::U32>()) {
TINT_ICE(Transform, b.Diagnostics())
@@ -121,12 +121,12 @@ struct Robustness::State {
b.Diagnostics().add_error(diag::System::Transform,
"unsupported constant value for accessor: " +
idx_constant.Type()->type_name(),
expr->source());
expr->source);
return nullptr;
}
} else {
// Dynamic value index
idx.expr = ctx.Clone(expr->idx_expr());
idx.expr = ctx.Clone(expr->index);
idx.is_signed = idx_ty->Is<sem::I32>();
}
@@ -178,8 +178,8 @@ struct Robustness::State {
}
// Clone arguments outside of create() call to have deterministic ordering
auto src = ctx.Clone(expr->source());
auto* arr = ctx.Clone(expr->array());
auto src = ctx.Clone(expr->source);
auto* arr = ctx.Clone(expr->array);
return b.IndexAccessor(src, arr, idx.expr);
}
@@ -214,8 +214,8 @@ struct Robustness::State {
auto array_idx = signature.IndexOf(sem::ParameterUsage::kArrayIndex);
auto level_idx = signature.IndexOf(sem::ParameterUsage::kLevel);
auto* texture_arg = expr->args()[texture_idx];
auto* coords_arg = expr->args()[coords_idx];
auto* texture_arg = expr->args[texture_idx];
auto* coords_arg = expr->args[coords_idx];
auto* coords_ty = intrinsic->Parameters()[coords_idx]->Type();
// If the level is provided, then we need to clamp this. As the level is
@@ -226,7 +226,7 @@ struct Robustness::State {
std::function<ast::Expression*()> level_arg;
if (level_idx >= 0) {
level_arg = [&] {
auto* arg = expr->args()[level_idx];
auto* arg = expr->args[level_idx];
auto* num_levels = b.Call("textureNumLevels", ctx.Clone(texture_arg));
auto* zero = b.Expr(0);
auto* max = ctx.dst->Sub(num_levels, 1);
@@ -250,7 +250,7 @@ struct Robustness::State {
// Clamp the array_index argument, if provided
if (array_idx >= 0) {
auto* arg = expr->args()[array_idx];
auto* arg = expr->args[array_idx];
auto* num_layers = b.Call("textureNumLayers", ctx.Clone(texture_arg));
auto* zero = b.Expr(0);
auto* max = ctx.dst->Sub(num_layers, 1);
@@ -260,7 +260,7 @@ struct Robustness::State {
// Clamp the level argument, if provided
if (level_idx >= 0) {
auto* arg = expr->args()[level_idx];
auto* arg = expr->args[level_idx];
ctx.Replace(arg, level_arg ? level_arg() : ctx.dst->Expr(0));
}

14
src/transform/simplify.cc
View File

@@ -37,16 +37,16 @@ Simplify::~Simplify() = default;
void Simplify::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ctx.ReplaceAll([&](ast::Expression* expr) -> ast::Expression* {
if (auto* outer = expr->As<ast::UnaryOpExpression>()) {
if (auto* inner = outer->expr()->As<ast::UnaryOpExpression>()) {
if (outer->op() == ast::UnaryOp::kAddressOf &&
inner->op() == ast::UnaryOp::kIndirection) {
if (auto* inner = outer->expr->As<ast::UnaryOpExpression>()) {
if (outer->op == ast::UnaryOp::kAddressOf &&
inner->op == ast::UnaryOp::kIndirection) {
// &(*(expr)) => expr
return ctx.Clone(inner->expr());
return ctx.Clone(inner->expr);
}
if (outer->op() == ast::UnaryOp::kIndirection &&
inner->op() == ast::UnaryOp::kAddressOf) {
if (outer->op == ast::UnaryOp::kIndirection &&
inner->op == ast::UnaryOp::kAddressOf) {
// *(&(expr)) => expr
return ctx.Clone(inner->expr());
return ctx.Clone(inner->expr);
}
}
}

6
src/transform/single_entry_point.cc
View File

@@ -47,7 +47,7 @@ void SingleEntryPoint::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
if (!f->IsEntryPoint()) {
continue;
}
if (ctx.src->Symbols().NameFor(f->symbol()) == cfg->entry_point_name) {
if (ctx.src->Symbols().NameFor(f->symbol) == cfg->entry_point_name) {
entry_point = f;
break;
}
@@ -74,11 +74,11 @@ void SingleEntryPoint::Run(CloneContext& ctx, const DataMap& inputs, DataMap&) {
// TODO(jrprice): Strip unused types.
ctx.dst->AST().AddTypeDecl(ctx.Clone(ty));
} else if (auto* var = decl->As<ast::Variable>()) {
if (var->is_const() || referenced_vars.count(var)) {
if (var->is_const || referenced_vars.count(var)) {
ctx.dst->AST().AddGlobalVariable(ctx.Clone(var));
}
} else if (auto* func = decl->As<ast::Function>()) {
if (sem.Get(func)->HasAncestorEntryPoint(entry_point->symbol())) {
if (sem.Get(func)->HasAncestorEntryPoint(entry_point->symbol)) {
ctx.dst->AST().AddFunction(ctx.Clone(func));
}
} else {

4
src/transform/transform.cc
View File

@@ -78,7 +78,7 @@ bool Transform::Requires(CloneContext& ctx,
void Transform::RemoveStatement(CloneContext& ctx, ast::Statement* stmt) {
auto* sem = ctx.src->Sem().Get(stmt);
if (auto* block = tint::As<sem::BlockStatement>(sem->Parent())) {
ctx.Remove(block->Declaration()->statements(), stmt);
ctx.Remove(block->Declaration()->statements, stmt);
return;
}
if (tint::Is<sem::ForLoopStatement>(sem->Parent())) {
@@ -127,7 +127,7 @@ ast::Type* Transform::CreateASTTypeFor(CloneContext& ctx, const sem::Type* ty) {
}
}
if (auto* s = ty->As<sem::Struct>()) {
return ctx.dst->create<ast::TypeName>(ctx.Clone(s->Declaration()->name()));
return ctx.dst->create<ast::TypeName>(ctx.Clone(s->Declaration()->name));
}
if (auto* s = ty->As<sem::Reference>()) {
return CreateASTTypeFor(ctx, s->StoreType());

51
src/transform/transform_test.cc
View File

@@ -62,9 +62,9 @@ TEST_F(CreateASTTypeForTest, Matrix) {
return b.create<sem::Matrix>(column_type, 3u);
});
ASSERT_TRUE(mat->Is<ast::Matrix>());
ASSERT_TRUE(mat->As<ast::Matrix>()->type()->Is<ast::F32>());
ASSERT_EQ(mat->As<ast::Matrix>()->columns(), 3u);
ASSERT_EQ(mat->As<ast::Matrix>()->rows(), 2u);
ASSERT_TRUE(mat->As<ast::Matrix>()->type->Is<ast::F32>());
ASSERT_EQ(mat->As<ast::Matrix>()->columns, 3u);
ASSERT_EQ(mat->As<ast::Matrix>()->rows, 2u);
}
TEST_F(CreateASTTypeForTest, Vector) {
@@ -72,8 +72,8 @@ TEST_F(CreateASTTypeForTest, Vector) {
return b.create<sem::Vector>(b.create<sem::F32>(), 2);
});
ASSERT_TRUE(vec->Is<ast::Vector>());
ASSERT_TRUE(vec->As<ast::Vector>()->type()->Is<ast::F32>());
ASSERT_EQ(vec->As<ast::Vector>()->size(), 2u);
ASSERT_TRUE(vec->As<ast::Vector>()->type->Is<ast::F32>());
ASSERT_EQ(vec->As<ast::Vector>()->width, 2u);
}
TEST_F(CreateASTTypeForTest, ArrayImplicitStride) {
@@ -81,15 +81,15 @@ TEST_F(CreateASTTypeForTest, ArrayImplicitStride) {
return b.create<sem::Array>(b.create<sem::F32>(), 2, 4, 4, 32u, 32u);
});
ASSERT_TRUE(arr->Is<ast::Array>());
ASSERT_TRUE(arr->As<ast::Array>()->type()->Is<ast::F32>());
ASSERT_EQ(arr->As<ast::Array>()->decorations().size(), 0u);
ASSERT_TRUE(arr->As<ast::Array>()->type->Is<ast::F32>());
ASSERT_EQ(arr->As<ast::Array>()->decorations.size(), 0u);
auto* size_expr =
arr->As<ast::Array>()->Size()->As<ast::ScalarConstructorExpression>();
ASSERT_NE(size_expr, nullptr);
auto* size = size_expr->literal()->As<ast::IntLiteral>();
auto* count_expr =
arr->As<ast::Array>()->count->As<ast::ScalarConstructorExpression>();
ASSERT_NE(count_expr, nullptr);
auto* size = count_expr->literal->As<ast::IntLiteral>();
ASSERT_NE(size, nullptr);
EXPECT_EQ(size->value_as_i32(), 2);
EXPECT_EQ(size->ValueAsI32(), 2);
}
TEST_F(CreateASTTypeForTest, ArrayNonImplicitStride) {
@@ -97,35 +97,34 @@ TEST_F(CreateASTTypeForTest, ArrayNonImplicitStride) {
return b.create<sem::Array>(b.create<sem::F32>(), 2, 4, 4, 64u, 32u);
});
ASSERT_TRUE(arr->Is<ast::Array>());
ASSERT_TRUE(arr->As<ast::Array>()->type()->Is<ast::F32>());
ASSERT_EQ(arr->As<ast::Array>()->decorations().size(), 1u);
ASSERT_TRUE(arr->As<ast::Array>()->type->Is<ast::F32>());
ASSERT_EQ(arr->As<ast::Array>()->decorations.size(), 1u);
ASSERT_TRUE(
arr->As<ast::Array>()->decorations()[0]->Is<ast::StrideDecoration>());
arr->As<ast::Array>()->decorations[0]->Is<ast::StrideDecoration>());
ASSERT_EQ(arr->As<ast::Array>()
->decorations()[0]
->decorations[0]
->As<ast::StrideDecoration>()
->stride(),
->stride,
64u);
auto* size_expr =
arr->As<ast::Array>()->Size()->As<ast::ScalarConstructorExpression>();
ASSERT_NE(size_expr, nullptr);
auto* size = size_expr->literal()->As<ast::IntLiteral>();
auto* count_expr =
arr->As<ast::Array>()->count->As<ast::ScalarConstructorExpression>();
ASSERT_NE(count_expr, nullptr);
auto* size = count_expr->literal->As<ast::IntLiteral>();
ASSERT_NE(size, nullptr);
EXPECT_EQ(size->value_as_i32(), 2);
EXPECT_EQ(size->ValueAsI32(), 2);
}
TEST_F(CreateASTTypeForTest, Struct) {
auto* str = create([](ProgramBuilder& b) {
auto* decl = b.Structure("S", {}, {});
return b.create<sem::Struct>(decl, decl->name(), sem::StructMemberList{},
return b.create<sem::Struct>(decl, decl->name, sem::StructMemberList{},
4 /* align */, 4 /* size */,
4 /* size_no_padding */);
});
ASSERT_TRUE(str->Is<ast::TypeName>());
EXPECT_EQ(
ast_type_builder.Symbols().NameFor(str->As<ast::TypeName>()->name()),
"S");
EXPECT_EQ(ast_type_builder.Symbols().NameFor(str->As<ast::TypeName>()->name),
"S");
}
} // namespace

68
src/transform/vertex_pulling.cc
View File

@@ -726,27 +726,27 @@ struct State {
/// @param param the parameter to process
void ProcessNonStructParameter(ast::Function* func, ast::Variable* param) {
if (auto* location =
ast::GetDecoration<ast::LocationDecoration>(param->decorations())) {
ast::GetDecoration<ast::LocationDecoration>(param->decorations)) {
// Create a function-scope variable to replace the parameter.
auto func_var_sym = ctx.Clone(param->symbol());
auto* func_var_type = ctx.Clone(param->type());
auto func_var_sym = ctx.Clone(param->symbol);
auto* func_var_type = ctx.Clone(param->type);
auto* func_var = ctx.dst->Var(func_var_sym, func_var_type);
ctx.InsertFront(func->body()->statements(), ctx.dst->Decl(func_var));
ctx.InsertFront(func->body->statements, ctx.dst->Decl(func_var));
// Capture mapping from location to the new variable.
LocationInfo info;
info.expr = [this, func_var]() { return ctx.dst->Expr(func_var); };
info.type = ctx.src->Sem().Get(param)->Type();
location_info[location->value()] = info;
location_info[location->value] = info;
} else if (auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(
param->decorations())) {
param->decorations)) {
// Check for existing vertex_index and instance_index builtins.
if (builtin->value() == ast::Builtin::kVertexIndex) {
if (builtin->builtin == ast::Builtin::kVertexIndex) {
vertex_index_expr = [this, param]() {
return ctx.dst->Expr(ctx.Clone(param->symbol()));
return ctx.dst->Expr(ctx.Clone(param->symbol));
};
} else if (builtin->value() == ast::Builtin::kInstanceIndex) {
} else if (builtin->builtin == ast::Builtin::kInstanceIndex) {
instance_index_expr = [this, param]() {
return ctx.dst->Expr(ctx.Clone(param->symbol()));
return ctx.dst->Expr(ctx.Clone(param->symbol));
};
}
new_function_parameters.push_back(ctx.Clone(param));
@@ -767,32 +767,32 @@ struct State {
void ProcessStructParameter(ast::Function* func,
ast::Variable* param,
const ast::Struct* struct_ty) {
auto param_sym = ctx.Clone(param->symbol());
auto param_sym = ctx.Clone(param->symbol);
// Process the struct members.
bool has_locations = false;
ast::StructMemberList members_to_clone;
for (auto* member : struct_ty->members()) {
auto member_sym = ctx.Clone(member->symbol());
for (auto* member : struct_ty->members) {
auto member_sym = ctx.Clone(member->symbol);
std::function<ast::Expression*()> member_expr = [this, param_sym,
member_sym]() {
return ctx.dst->MemberAccessor(param_sym, member_sym);
};
if (auto* location = ast::GetDecoration<ast::LocationDecoration>(
member->decorations())) {
member->decorations)) {
// Capture mapping from location to struct member.
LocationInfo info;
info.expr = member_expr;
info.type = ctx.src->Sem().Get(member)->Type();
location_info[location->value()] = info;
location_info[location->value] = info;
has_locations = true;
} else if (auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(
member->decorations())) {
member->decorations)) {
// Check for existing vertex_index and instance_index builtins.
if (builtin->value() == ast::Builtin::kVertexIndex) {
if (builtin->builtin == ast::Builtin::kVertexIndex) {
vertex_index_expr = member_expr;
} else if (builtin->value() == ast::Builtin::kInstanceIndex) {
} else if (builtin->builtin == ast::Builtin::kInstanceIndex) {
instance_index_expr = member_expr;
}
members_to_clone.push_back(member);
@@ -809,16 +809,16 @@ struct State {
}
// Create a function-scope variable to replace the parameter.
auto* func_var = ctx.dst->Var(param_sym, ctx.Clone(param->type()));
ctx.InsertFront(func->body()->statements(), ctx.dst->Decl(func_var));
auto* func_var = ctx.dst->Var(param_sym, ctx.Clone(param->type));
ctx.InsertFront(func->body->statements, ctx.dst->Decl(func_var));
if (!members_to_clone.empty()) {
// Create a new struct without the location attributes.
ast::StructMemberList new_members;
for (auto* member : members_to_clone) {
auto member_sym = ctx.Clone(member->symbol());
auto* member_type = ctx.Clone(member->type());
auto member_decos = ctx.Clone(member->decorations());
auto member_sym = ctx.Clone(member->symbol);
auto* member_type = ctx.Clone(member->type);
auto member_decos = ctx.Clone(member->decorations);
new_members.push_back(
ctx.dst->Member(member_sym, member_type, std::move(member_decos)));
}
@@ -831,9 +831,9 @@ struct State {
// Copy values from the new parameter to the function-scope variable.
for (auto* member : members_to_clone) {
auto member_name = ctx.Clone(member->symbol());
auto member_name = ctx.Clone(member->symbol);
ctx.InsertFront(
func->body()->statements(),
func->body->statements,
ctx.dst->Assign(ctx.dst->MemberAccessor(func_var, member_name),
ctx.dst->MemberAccessor(new_param, member_name)));
}
@@ -843,12 +843,12 @@ struct State {
/// Process an entry point function.
/// @param func the entry point function
void Process(ast::Function* func) {
if (func->body()->empty()) {
if (func->body->Empty()) {
return;
}
// Process entry point parameters.
for (auto* param : func->params()) {
for (auto* param : func->params) {
auto* sem = ctx.src->Sem().Get(param);
if (auto* str = sem->Type()->As<sem::Struct>()) {
ProcessStructParameter(func, param, str->Declaration());
@@ -885,17 +885,17 @@ struct State {
// Generate vertex pulling preamble.
if (auto* block = CreateVertexPullingPreamble()) {
ctx.InsertFront(func->body()->statements(), block);
ctx.InsertFront(func->body->statements, block);
}
// Rewrite the function header with the new parameters.
auto func_sym = ctx.Clone(func->symbol());
auto* ret_type = ctx.Clone(func->return_type());
auto* body = ctx.Clone(func->body());
auto decos = ctx.Clone(func->decorations());
auto ret_decos = ctx.Clone(func->return_type_decorations());
auto func_sym = ctx.Clone(func->symbol);
auto* ret_type = ctx.Clone(func->return_type);
auto* body = ctx.Clone(func->body);
auto decos = ctx.Clone(func->decorations);
auto ret_decos = ctx.Clone(func->return_type_decorations);
auto* new_func = ctx.dst->create<ast::Function>(
func->source(), func_sym, new_function_parameters, ret_type, body,
func->source, func_sym, new_function_parameters, ret_type, body,
std::move(decos), std::move(ret_decos));
ctx.Replace(func, new_func);
}

13
src/transform/wrap_arrays_in_structs.cc
View File

@@ -60,14 +60,14 @@ void WrapArraysInStructs::Run(CloneContext& ctx, const DataMap&, DataMap&) {
ctx.ReplaceAll([&](ast::ArrayAccessorExpression* accessor)
-> ast::ArrayAccessorExpression* {
if (auto* array = ::tint::As<sem::Array>(
sem.Get(accessor->array())->Type()->UnwrapRef())) {
sem.Get(accessor->array)->Type()->UnwrapRef())) {
if (wrapper(array)) {
// Array is wrapped in a structure. Emit a member accessor to get
// to the actual array.
auto* arr = ctx.Clone(accessor->array());
auto* idx = ctx.Clone(accessor->idx_expr());
auto* arr = ctx.Clone(accessor->array);
auto* idx = ctx.Clone(accessor->index);
auto* unwrapped = ctx.dst->MemberAccessor(arr, "arr");
return ctx.dst->IndexAccessor(accessor->source(), unwrapped, idx);
return ctx.dst->IndexAccessor(accessor->source, unwrapped, idx);
}
}
return nullptr;
@@ -80,10 +80,9 @@ void WrapArraysInStructs::Run(CloneContext& ctx, const DataMap&, DataMap&) {
if (auto w = wrapper(array)) {
// Wrap the array type constructor with another constructor for
// the wrapper
auto* wrapped_array_ty = ctx.Clone(ctor->type());
auto* wrapped_array_ty = ctx.Clone(ctor->type);
auto* array_ty = w.array_type(ctx);
auto* arr_ctor =
ctx.dst->Construct(array_ty, ctx.Clone(ctor->values()));
auto* arr_ctor = ctx.dst->Construct(array_ty, ctx.Clone(ctor->values));
return ctx.dst->Construct(wrapped_array_ty, arr_ctor);
}
}

40
src/transform/zero_init_workgroup_memory.cc
View File

@@ -116,7 +116,7 @@ struct ZeroInitWorkgroupMemory::State {
auto& sem = ctx.src->Sem();
CalculateWorkgroupSize(
ast::GetDecoration<ast::WorkgroupDecoration>(fn->decorations()));
ast::GetDecoration<ast::WorkgroupDecoration>(fn->decorations));
// Generate a list of statements to zero initialize each of the
// workgroup storage variables used by `fn`. This will populate #statements.
@@ -125,7 +125,7 @@ struct ZeroInitWorkgroupMemory::State {
if (var->StorageClass() == ast::StorageClass::kWorkgroup) {
BuildZeroingStatements(
var->Type()->UnwrapRef(), [&](uint32_t num_values) {
auto var_name = ctx.Clone(var->Declaration()->symbol());
auto var_name = ctx.Clone(var->Declaration()->symbol);
return Expression{b.Expr(var_name), num_values, ArrayIndices{}};
});
}
@@ -138,11 +138,11 @@ struct ZeroInitWorkgroupMemory::State {
// Scan the entry point for an existing local_invocation_index builtin
// parameter
std::function<ast::Expression*()> local_index;
for (auto* param : fn->params()) {
if (auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(
param->decorations())) {
if (builtin->value() == ast::Builtin::kLocalInvocationIndex) {
local_index = [=] { return b.Expr(ctx.Clone(param->symbol())); };
for (auto* param : fn->params) {
if (auto* builtin =
ast::GetDecoration<ast::BuiltinDecoration>(param->decorations)) {
if (builtin->builtin == ast::Builtin::kLocalInvocationIndex) {
local_index = [=] { return b.Expr(ctx.Clone(param->symbol)); };
break;
}
}
@@ -150,11 +150,11 @@ struct ZeroInitWorkgroupMemory::State {
if (auto* str = sem.Get(param)->Type()->As<sem::Struct>()) {
for (auto* member : str->Members()) {
if (auto* builtin = ast::GetDecoration<ast::BuiltinDecoration>(
member->Declaration()->decorations())) {
if (builtin->value() == ast::Builtin::kLocalInvocationIndex) {
member->Declaration()->decorations)) {
if (builtin->builtin == ast::Builtin::kLocalInvocationIndex) {
local_index = [=] {
auto* param_expr = b.Expr(ctx.Clone(param->symbol()));
auto member_name = ctx.Clone(member->Declaration()->symbol());
auto* param_expr = b.Expr(ctx.Clone(param->symbol));
auto member_name = ctx.Clone(member->Declaration()->symbol);
return b.MemberAccessor(param_expr, member_name);
};
break;
@@ -168,8 +168,8 @@ struct ZeroInitWorkgroupMemory::State {
auto* param =
b.Param(b.Symbols().New("local_invocation_index"), b.ty.u32(),
{b.Builtin(ast::Builtin::kLocalInvocationIndex)});
ctx.InsertBack(fn->params(), param);
local_index = [=] { return b.Expr(param->symbol()); };
ctx.InsertBack(fn->params, param);
local_index = [=] { return b.Expr(param->symbol); };
}
// Take the zeroing statements and bin them by the number of iterations
@@ -225,7 +225,7 @@ struct ZeroInitWorkgroupMemory::State {
block.emplace_back(s.stmt);
}
auto* for_loop = b.For(init, cond, cont, b.Block(block));
ctx.InsertFront(fn->body()->statements(), for_loop);
ctx.InsertFront(fn->body->statements, for_loop);
} else if (num_iterations < workgroup_size_const) {
// Workgroup size is a known constant, but is greater than
// num_iterations. Emit an if statement:
@@ -241,7 +241,7 @@ struct ZeroInitWorkgroupMemory::State {
block.emplace_back(s.stmt);
}
auto* if_stmt = b.If(cond, b.Block(block));
ctx.InsertFront(fn->body()->statements(), if_stmt);
ctx.InsertFront(fn->body->statements, if_stmt);
} else {
// Workgroup size exactly equals num_iterations.
// No need for any conditionals. Just emit a basic block:
@@ -254,12 +254,12 @@ struct ZeroInitWorkgroupMemory::State {
for (auto& s : stmts) {
block.emplace_back(s.stmt);
}
ctx.InsertFront(fn->body()->statements(), b.Block(block));
ctx.InsertFront(fn->body->statements, b.Block(block));
}
}
// Append a single workgroup barrier after the zero initialization.
ctx.InsertFront(fn->body()->statements(),
ctx.InsertFront(fn->body->statements,
b.create<ast::CallStatement>(b.Call("workgroupBarrier")));
}
@@ -294,7 +294,7 @@ struct ZeroInitWorkgroupMemory::State {
if (auto* str = ty->As<sem::Struct>()) {
for (auto* member : str->Members()) {
auto name = ctx.Clone(member->Declaration()->symbol());
auto name = ctx.Clone(member->Declaration()->symbol);
BuildZeroingStatements(member->Type(), [&](uint32_t num_values) {
auto s = get_expr(num_values);
return Expression{b.MemberAccessor(s.expr, name), s.num_iterations,
@@ -365,7 +365,7 @@ struct ZeroInitWorkgroupMemory::State {
bool is_signed = false;
workgroup_size_const = 1u;
workgroup_size_expr = nullptr;
for (auto* expr : deco->values()) {
for (auto* expr : deco->Values()) {
if (!expr) {
continue;
}
@@ -436,7 +436,7 @@ ZeroInitWorkgroupMemory::~ZeroInitWorkgroupMemory() = default;
void ZeroInitWorkgroupMemory::Run(CloneContext& ctx, const DataMap&, DataMap&) {
for (auto* fn : ctx.src->AST().Functions()) {
if (fn->pipeline_stage() == ast::PipelineStage::kCompute) {
if (fn->PipelineStage() == ast::PipelineStage::kCompute) {
State{ctx}.Run(fn);
}
}