writer/msl: Rework string printing

Add `out` parameters to expression and type generators.

Use the new helper classes in TextGenerator.
Cleans up bad formatting.

Prepares the writer generating 'pre' statements, required for atomics.

If-else statements are generated slightly differently. This is done so that 'pre' statements for the else conditions are scoped correctly. This is identical to the HLSL writer.

Bug tint:892

Change-Id: I4c6e96c90673ba30898b3682bf3198497d63a2d4
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/56067
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>

src/writer/msl/generator_impl.h

@@ -64,62 +64,63 @@ class GeneratorImpl : public TextGenerator {
   /// @returns true if the declared type was emitted
   bool EmitTypeDecl(const sem::Type* ty);
   /// Handles an array accessor expression
+  /// @param out the output of the expression stream
   /// @param expr the expression to emit
   /// @returns true if the array accessor was emitted
-  bool EmitArrayAccessor(ast::ArrayAccessorExpression* expr);
+  bool EmitArrayAccessor(std::ostream& out, ast::ArrayAccessorExpression* expr);
   /// Handles an assignment statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted successfully
   bool EmitAssign(ast::AssignmentStatement* stmt);
   /// Handles generating a binary expression
+  /// @param out the output of the expression stream
   /// @param expr the binary expression
   /// @returns true if the expression was emitted, false otherwise
-  bool EmitBinary(ast::BinaryExpression* expr);
+  bool EmitBinary(std::ostream& out, ast::BinaryExpression* expr);
   /// Handles generating a bitcast expression
+  /// @param out the output of the expression stream
   /// @param expr the bitcast expression
   /// @returns true if the bitcast was emitted
-  bool EmitBitcast(ast::BitcastExpression* expr);
+  bool EmitBitcast(std::ostream& out, ast::BitcastExpression* expr);
   /// Handles a block statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted successfully
   bool EmitBlock(const ast::BlockStatement* stmt);
-  /// Handles a block statement with a newline at the end
-  /// @param stmt the statement to emit
-  /// @returns true if the statement was emitted successfully
-  bool EmitIndentedBlockAndNewline(ast::BlockStatement* stmt);
-  /// Handles a block statement with a newline at the end
-  /// @param stmt the statement to emit
-  /// @returns true if the statement was emitted successfully
-  bool EmitBlockAndNewline(const ast::BlockStatement* stmt);
   /// Handles a break statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted successfully
   bool EmitBreak(ast::BreakStatement* stmt);
   /// Handles generating a call expression
+  /// @param out the output of the expression stream
   /// @param expr the call expression
   /// @returns true if the call expression is emitted
-  bool EmitCall(ast::CallExpression* expr);
+  bool EmitCall(std::ostream& out, ast::CallExpression* expr);
   /// Handles generating an intrinsic call expression
+  /// @param out the output of the expression stream
   /// @param expr the call expression
   /// @param intrinsic the intrinsic being called
   /// @returns true if the call expression is emitted
-  bool EmitIntrinsicCall(ast::CallExpression* expr,
+  bool EmitIntrinsicCall(std::ostream& out,
+                         ast::CallExpression* expr,
                          const sem::Intrinsic* intrinsic);
   /// Handles generating a call to a texture function (`textureSample`,
   /// `textureSampleGrad`, etc)
+  /// @param out the output of the expression stream
   /// @param expr the call expression
   /// @param intrinsic the semantic information for the texture intrinsic
   /// @returns true if the call expression is emitted
-  bool EmitTextureCall(ast::CallExpression* expr,
+  bool EmitTextureCall(std::ostream& out,
+                       ast::CallExpression* expr,
                        const sem::Intrinsic* intrinsic);
   /// Handles a case statement
   /// @param stmt the statement
   /// @returns true if the statement was emitted successfully
   bool EmitCase(ast::CaseStatement* stmt);
   /// Handles generating constructor expressions
+  /// @param out the output of the expression stream
   /// @param expr the constructor expression
   /// @returns true if the expression was emitted
-  bool EmitConstructor(ast::ConstructorExpression* expr);
+  bool EmitConstructor(std::ostream& out, ast::ConstructorExpression* expr);
   /// Handles a continue statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted successfully
@@ -128,85 +129,106 @@ class GeneratorImpl : public TextGenerator {
   /// @param stmt the discard statement
   /// @returns true if the statement was successfully emitted
   bool EmitDiscard(ast::DiscardStatement* stmt);
-  /// Handles generating an else statement
-  /// @param stmt the statement to emit
-  /// @returns true if the statement was emitted
-  bool EmitElse(ast::ElseStatement* stmt);
   /// Handles emitting the entry point function
   /// @param func the entry point function
   /// @returns true if the entry point function was emitted
   bool EmitEntryPointFunction(ast::Function* func);
   /// Handles generate an Expression
+  /// @param out the output of the expression stream
   /// @param expr the expression
   /// @returns true if the expression was emitted
-  bool EmitExpression(ast::Expression* expr);
+  bool EmitExpression(std::ostream& out, ast::Expression* expr);
   /// Handles generating a function
   /// @param func the function to generate
   /// @returns true if the function was emitted
   bool EmitFunction(ast::Function* func);
   /// Handles generating an identifier expression
+  /// @param out the output of the expression stream
   /// @param expr the identifier expression
   /// @returns true if the identifier was emitted
-  bool EmitIdentifier(ast::IdentifierExpression* expr);
+  bool EmitIdentifier(std::ostream& out, ast::IdentifierExpression* expr);
   /// Handles an if statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was successfully emitted
   bool EmitIf(ast::IfStatement* stmt);
   /// Handles a literal
+  /// @param out the output of the expression stream
   /// @param lit the literal to emit
   /// @returns true if the literal was successfully emitted
-  bool EmitLiteral(ast::Literal* lit);
+  bool EmitLiteral(std::ostream& out, ast::Literal* lit);
   /// Handles a loop statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted
   bool EmitLoop(ast::LoopStatement* stmt);
   /// Handles a member accessor expression
+  /// @param out the output of the expression stream
   /// @param expr the member accessor expression
   /// @returns true if the member accessor was emitted
-  bool EmitMemberAccessor(ast::MemberAccessorExpression* expr);
+  bool EmitMemberAccessor(std::ostream& out,
+                          ast::MemberAccessorExpression* expr);
   /// Handles return statements
   /// @param stmt the statement to emit
   /// @returns true if the statement was successfully emitted
   bool EmitReturn(ast::ReturnStatement* stmt);
   /// Handles generating a scalar constructor
+  /// @param out the output of the expression stream
   /// @param expr the scalar constructor expression
   /// @returns true if the scalar constructor is emitted
-  bool EmitScalarConstructor(ast::ScalarConstructorExpression* expr);
+  bool EmitScalarConstructor(std::ostream& out,
+                             ast::ScalarConstructorExpression* expr);
   /// Handles emitting a pipeline stage name
+  /// @param out the output of the expression stream
   /// @param stage the stage to emit
-  void EmitStage(ast::PipelineStage stage);
+  void EmitStage(std::ostream& out, ast::PipelineStage stage);
   /// Handles statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted
   bool EmitStatement(ast::Statement* stmt);
+  /// Emits a list of statements
+  /// @param stmts the statement list
+  /// @returns true if the statements were emitted successfully
+  bool EmitStatements(const ast::StatementList& stmts);
+  /// Emits a list of statements with an indentation
+  /// @param stmts the statement list
+  /// @returns true if the statements were emitted successfully
+  bool EmitStatementsWithIndent(const ast::StatementList& stmts);
   /// Handles generating a switch statement
   /// @param stmt the statement to emit
   /// @returns true if the statement was emitted
   bool EmitSwitch(ast::SwitchStatement* stmt);
   /// Handles generating a type
+  /// @param out the output of the type stream
   /// @param type the type to generate
   /// @param name the name of the variable, only used for array emission
   /// @returns true if the type is emitted
-  bool EmitType(const sem::Type* type, const std::string& name);
+  bool EmitType(std::ostream& out,
+                const sem::Type* type,
+                const std::string& name);
   /// Handles generating an MSL-packed storage type.
   /// If the type does not have a packed form, the standard non-packed form is
   /// emitted.
+  /// @param out the output of the type stream
   /// @param type the type to generate
   /// @param name the name of the variable, only used for array emission
   /// @returns true if the type is emitted
-  bool EmitPackedType(const sem::Type* type, const std::string& name);
+  bool EmitPackedType(std::ostream& out,
+                      const sem::Type* type,
+                      const std::string& name);
   /// Handles generating a struct declaration
   /// @param str the struct to generate
   /// @returns true if the struct is emitted
   bool EmitStructType(const sem::Struct* str);
   /// Handles emitting a type constructor
+  /// @param out the output of the expression stream
   /// @param expr the type constructor expression
   /// @returns true if the constructor is emitted
-  bool EmitTypeConstructor(ast::TypeConstructorExpression* expr);
+  bool EmitTypeConstructor(std::ostream& out,
+                           ast::TypeConstructorExpression* expr);
   /// Handles a unary op expression
+  /// @param out the output of the expression stream
   /// @param expr the expression to emit
   /// @returns true if the expression was emitted
-  bool EmitUnaryOp(ast::UnaryOpExpression* expr);
+  bool EmitUnaryOp(std::ostream& out, ast::UnaryOpExpression* expr);
   /// Handles generating a variable
   /// @param var the variable to generate
   /// @returns true if the variable was emitted
@@ -216,9 +238,10 @@ class GeneratorImpl : public TextGenerator {
   /// @returns true if the variable was emitted
   bool EmitProgramConstVariable(const ast::Variable* var);
   /// Emits the zero value for the given type
+  /// @param out the output of the expression stream
   /// @param type the type to emit the value for
   /// @returns true if the zero value was successfully emitted.
-  bool EmitZeroValue(const sem::Type* type);
+  bool EmitZeroValue(std::ostream& out, const sem::Type* type);
 
   /// Handles generating a builtin name
   /// @param intrinsic the semantic info for the intrinsic
@@ -259,15 +282,6 @@ class GeneratorImpl : public TextGenerator {
   /// type.
   SizeAndAlign MslPackedTypeSizeAndAlign(const sem::Type* ty);
 
-  /// Emits `prefix`, followed by an opening brace `{`, then calls `cb` to emit
-  /// the block body, then finally emits the closing brace `}`.
-  /// @param prefix the string to emit before the opening brace
-  /// @param cb a function or function-like object with the signature `bool()`
-  /// that emits the block body.
-  /// @returns the return value of `cb`.
-  template <typename F>
-  bool EmitBlockBraces(const std::string& prefix, F&& cb);
-
   const Program* program_ = nullptr;
   std::function<bool()> emit_continuing_;
 };

src/writer/msl/generator_impl_array_accessor_test.cc

@@ -28,8 +28,9 @@ TEST_F(MslGeneratorImplTest, ArrayAccessor) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "ary[5]");
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), "ary[5]");
 }
 
 TEST_F(MslGeneratorImplTest, ArrayAccessor_OfDref) {
@@ -41,8 +42,9 @@ TEST_F(MslGeneratorImplTest, ArrayAccessor_OfDref) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "(*(p))[5]");
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), "(*(p))[5]");
 }
 
 }  // namespace

src/writer/msl/generator_impl_binary_test.cc

@@ -47,8 +47,9 @@ TEST_P(MslBinaryTest, Emit) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), params.result);
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), params.result);
 }
 INSTANTIATE_TEST_SUITE_P(
     MslGeneratorImplTest,

src/writer/msl/generator_impl_bitcast_test.cc

@@ -27,8 +27,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_Bitcast) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(bitcast)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "as_type<float>(1)");
+  ASSERT_TRUE(gen.EmitExpression(out, bitcast)) << gen.error();
+  EXPECT_EQ(out.str(), "as_type<float>(1)");
 }
 
 }  // namespace

src/writer/msl/generator_impl_block_test.cc

@@ -45,9 +45,10 @@ TEST_F(MslGeneratorImplTest, Emit_Block_WithoutNewline) {
   gen.increment_indent();
 
   ASSERT_TRUE(gen.EmitBlock(b)) << gen.error();
-  EXPECT_EQ(gen.result(), R"({
+  EXPECT_EQ(gen.result(), R"(  {
     discard_fragment();
-  })");
+  }
+)");
 }
 
 }  // namespace

src/writer/msl/generator_impl_call_test.cc

@@ -31,8 +31,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_Call_WithoutParams) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "my_func()");
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
+  EXPECT_EQ(out.str(), "my_func()");
 }
 
 TEST_F(MslGeneratorImplTest, EmitExpression_Call_WithParams) {
@@ -50,8 +51,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_Call_WithParams) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "my_func(param1, param2)");
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
+  EXPECT_EQ(out.str(), "my_func(param1, param2)");
 }
 
 TEST_F(MslGeneratorImplTest, EmitStatement_Call) {

src/writer/msl/generator_impl_cast_test.cc

@@ -27,8 +27,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_Cast_Scalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(cast)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "float(1)");
+  ASSERT_TRUE(gen.EmitExpression(out, cast)) << gen.error();
+  EXPECT_EQ(out.str(), "float(1)");
 }
 
 TEST_F(MslGeneratorImplTest, EmitExpression_Cast_Vector) {
@@ -37,8 +38,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_Cast_Vector) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(cast)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "float3(int3(1, 2, 3))");
+  ASSERT_TRUE(gen.EmitExpression(out, cast)) << gen.error();
+  EXPECT_EQ(out.str(), "float3(int3(1, 2, 3))");
 }
 
 }  // namespace

src/writer/msl/generator_impl_function_test.cc

@@ -36,9 +36,9 @@ TEST_F(MslGeneratorImplTest, Emit_Function) {
   gen.increment_indent();
 
   ASSERT_TRUE(gen.Generate()) << gen.error();
-  EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
+  EXPECT_EQ(gen.result(), R"(  #include <metal_stdlib>
 
-using namespace metal;
+  using namespace metal;
   void my_func() {
     return;
   }
@@ -62,9 +62,9 @@ TEST_F(MslGeneratorImplTest, Emit_Function_WithParams) {
   gen.increment_indent();
 
   ASSERT_TRUE(gen.Generate()) << gen.error();
-  EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
+  EXPECT_EQ(gen.result(), R"(  #include <metal_stdlib>
 
-using namespace metal;
+  using namespace metal;
   void my_func(float a, int b) {
     return;
   }
@@ -564,9 +564,9 @@ TEST_F(MslGeneratorImplTest, Emit_Function_WithArrayParams) {
   gen.increment_indent();
 
   ASSERT_TRUE(gen.Generate()) << gen.error();
-  EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
+  EXPECT_EQ(gen.result(), R"(  #include <metal_stdlib>
 
-using namespace metal;
+  using namespace metal;
   struct tint_array_wrapper {
     float arr[5];
   };
@@ -589,9 +589,9 @@ TEST_F(MslGeneratorImplTest, Emit_Function_WithArrayReturn) {
   gen.increment_indent();
 
   ASSERT_TRUE(gen.Generate()) << gen.error();
-  EXPECT_EQ(gen.result(), R"(#include <metal_stdlib>
+  EXPECT_EQ(gen.result(), R"(  #include <metal_stdlib>
 
-using namespace metal;
+  using namespace metal;
   struct tint_array_wrapper {
     float arr[5];
   };

src/writer/msl/generator_impl_identifier_test.cc

@@ -29,8 +29,9 @@ TEST_F(MslGeneratorImplTest, EmitIdentifierExpression) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(i)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "foo");
+  ASSERT_TRUE(gen.EmitExpression(out, i)) << gen.error();
+  EXPECT_EQ(out.str(), "foo");
 }
 
 }  // namespace

src/writer/msl/generator_impl_if_test.cc

@@ -50,8 +50,10 @@ TEST_F(MslGeneratorImplTest, Emit_IfWithElseIf) {
   ASSERT_TRUE(gen.EmitStatement(i)) << gen.error();
   EXPECT_EQ(gen.result(), R"(  if (cond) {
     return;
-  } else if (else_cond) {
+  } else {
-    return;
+    if (else_cond) {
+      return;
+    }
   }
 )");
 }
@@ -88,10 +90,12 @@ TEST_F(MslGeneratorImplTest, Emit_IfWithMultiple) {
   ASSERT_TRUE(gen.EmitStatement(i)) << gen.error();
   EXPECT_EQ(gen.result(), R"(  if (cond) {
     return;
-  } else if (else_cond) {
-    return;
   } else {
-    return;
+    if (else_cond) {
+      return;
+    } else {
+      return;
+    }
   }
 )");
 }

src/writer/msl/generator_impl_import_test.cc

@@ -81,8 +81,9 @@ TEST_F(MslGeneratorImplTest, MslImportData_SingleParamTest_IntScalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), R"(abs(1))");
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), R"(abs(1))");
 }
 
 using MslImportData_DualParamTest = TestParamHelper<MslImportData>;
@@ -94,8 +95,9 @@ TEST_P(MslImportData_DualParamTest, FloatScalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string(param.msl_name) + "(1.0f, 2.0f)");
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), std::string(param.msl_name) + "(1.0f, 2.0f)");
 }
 INSTANTIATE_TEST_SUITE_P(MslGeneratorImplTest,
                          MslImportData_DualParamTest,
@@ -117,10 +119,11 @@ TEST_P(MslImportData_DualParam_VectorTest, FloatVector) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string(param.msl_name) +
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
-                              "(float3(1.0f, 2.0f, 3.0f), "
+  EXPECT_EQ(out.str(), std::string(param.msl_name) +
-                              "float3(4.0f, 5.0f, 6.0f))");
+                           "(float3(1.0f, 2.0f, 3.0f), "
+                           "float3(4.0f, 5.0f, 6.0f))");
 }
 INSTANTIATE_TEST_SUITE_P(MslGeneratorImplTest,
                          MslImportData_DualParam_VectorTest,
@@ -135,8 +138,9 @@ TEST_P(MslImportData_DualParam_Int_Test, IntScalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string(param.msl_name) + "(1, 2)");
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), std::string(param.msl_name) + "(1, 2)");
 }
 INSTANTIATE_TEST_SUITE_P(MslGeneratorImplTest,
                          MslImportData_DualParam_Int_Test,
@@ -152,8 +156,9 @@ TEST_P(MslImportData_TripleParamTest, FloatScalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string(param.msl_name) + "(1.0f, 2.0f, 3.0f)");
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), std::string(param.msl_name) + "(1.0f, 2.0f, 3.0f)");
 }
 INSTANTIATE_TEST_SUITE_P(
     MslGeneratorImplTest,
@@ -173,8 +178,9 @@ TEST_P(MslImportData_TripleParam_Int_Test, IntScalar) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string(param.msl_name) + "(1, 2, 3)");
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), std::string(param.msl_name) + "(1, 2, 3)");
 }
 INSTANTIATE_TEST_SUITE_P(MslGeneratorImplTest,
                          MslImportData_TripleParam_Int_Test,
@@ -190,8 +196,9 @@ TEST_F(MslGeneratorImplTest, MslImportData_Determinant) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitCall(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), std::string("determinant(var)"));
+  ASSERT_TRUE(gen.EmitCall(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), std::string("determinant(var)"));
 }
 
 }  // namespace

src/writer/msl/generator_impl_intrinsic_test.cc

@@ -291,9 +291,9 @@ TEST_F(MslGeneratorImplTest, Intrinsic_Call) {
 
   GeneratorImpl& gen = Build();
 
-  gen.increment_indent();
+  std::stringstream out;
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
-  EXPECT_EQ(gen.result(), "  dot(param1, param2)");
+  EXPECT_EQ(out.str(), "dot(param1, param2)");
 }
 
 TEST_F(MslGeneratorImplTest, StorageBarrier) {
@@ -302,9 +302,9 @@ TEST_F(MslGeneratorImplTest, StorageBarrier) {
 
   GeneratorImpl& gen = Build();
 
-  gen.increment_indent();
+  std::stringstream out;
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
-  EXPECT_EQ(gen.result(), "  threadgroup_barrier(mem_flags::mem_device)");
+  EXPECT_EQ(out.str(), "threadgroup_barrier(mem_flags::mem_device)");
 }
 
 TEST_F(MslGeneratorImplTest, WorkgroupBarrier) {
@@ -313,9 +313,9 @@ TEST_F(MslGeneratorImplTest, WorkgroupBarrier) {
 
   GeneratorImpl& gen = Build();
 
-  gen.increment_indent();
+  std::stringstream out;
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
-  EXPECT_EQ(gen.result(), "  threadgroup_barrier(mem_flags::mem_threadgroup)");
+  EXPECT_EQ(out.str(), "threadgroup_barrier(mem_flags::mem_threadgroup)");
 }
 
 TEST_F(MslGeneratorImplTest, Pack2x16Float) {
@@ -325,9 +325,9 @@ TEST_F(MslGeneratorImplTest, Pack2x16Float) {
 
   GeneratorImpl& gen = Build();
 
-  gen.increment_indent();
+  std::stringstream out;
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
-  EXPECT_EQ(gen.result(), "  as_type<uint>(half2(p1))");
+  EXPECT_EQ(out.str(), "as_type<uint>(half2(p1))");
 }
 
 TEST_F(MslGeneratorImplTest, Unpack2x16Float) {
@@ -337,9 +337,9 @@ TEST_F(MslGeneratorImplTest, Unpack2x16Float) {
 
   GeneratorImpl& gen = Build();
 
-  gen.increment_indent();
+  std::stringstream out;
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
-  EXPECT_EQ(gen.result(), "  float2(as_type<half2>(p1))");
+  EXPECT_EQ(out.str(), "float2(as_type<half2>(p1))");
 }
 
 TEST_F(MslGeneratorImplTest, Ignore) {

src/writer/msl/generator_impl_intrinsic_texture_test.cc

@@ -281,10 +281,11 @@ TEST_P(MslGeneratorIntrinsicTextureTest, Call) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(call)) << gen.error();
+  std::stringstream out;
+  ASSERT_TRUE(gen.EmitExpression(out, call)) << gen.error();
 
   auto expected = expected_texture_overload(param.overload);
-  EXPECT_EQ(expected, gen.result());
+  EXPECT_EQ(expected, out.str());
 }
 
 INSTANTIATE_TEST_SUITE_P(

src/writer/msl/generator_impl_member_accessor_test.cc

@@ -29,8 +29,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_MemberAccessor) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "str.mem");
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), "str.mem");
 }
 
 TEST_F(MslGeneratorImplTest, EmitExpression_MemberAccessor_Swizzle_xyz) {
@@ -40,8 +41,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_MemberAccessor_Swizzle_xyz) {
   WrapInFunction(expr);
 
   GeneratorImpl& gen = Build();
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "my_vec.xyz");
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), "my_vec.xyz");
 }
 
 TEST_F(MslGeneratorImplTest, EmitExpression_MemberAccessor_Swizzle_gbr) {
@@ -51,8 +53,9 @@ TEST_F(MslGeneratorImplTest, EmitExpression_MemberAccessor_Swizzle_gbr) {
   WrapInFunction(expr);
 
   GeneratorImpl& gen = Build();
-  ASSERT_TRUE(gen.EmitExpression(expr)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "my_vec.gbr");
+  ASSERT_TRUE(gen.EmitExpression(out, expr)) << gen.error();
+  EXPECT_EQ(out.str(), "my_vec.gbr");
 }
 
 }  // namespace

src/writer/msl/generator_impl_type_test.cc

@@ -66,8 +66,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Array) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(arr), "ary")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool ary[4]");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(arr), "ary")) << gen.error();
+  EXPECT_EQ(out.str(), "bool ary[4]");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_ArrayOfArray) {
@@ -77,8 +78,9 @@ TEST_F(MslGeneratorImplTest, EmitType_ArrayOfArray) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(b), "ary")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool ary[5][4]");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(b), "ary")) << gen.error();
+  EXPECT_EQ(out.str(), "bool ary[5][4]");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_ArrayOfArrayOfArray) {
@@ -89,8 +91,9 @@ TEST_F(MslGeneratorImplTest, EmitType_ArrayOfArrayOfArray) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(c), "ary")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool ary[6][5][4]");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(c), "ary")) << gen.error();
+  EXPECT_EQ(out.str(), "bool ary[6][5][4]");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Array_WithoutName) {
@@ -99,8 +102,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Array_WithoutName) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(arr), "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool[4]");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(arr), "")) << gen.error();
+  EXPECT_EQ(out.str(), "bool[4]");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_RuntimeArray) {
@@ -109,8 +113,9 @@ TEST_F(MslGeneratorImplTest, EmitType_RuntimeArray) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(arr), "ary")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool ary[1]");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(arr), "ary")) << gen.error();
+  EXPECT_EQ(out.str(), "bool ary[1]");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_ArrayWithStride) {
@@ -143,8 +148,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Bool) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(bool_, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "bool");
+  ASSERT_TRUE(gen.EmitType(out, bool_, "")) << gen.error();
+  EXPECT_EQ(out.str(), "bool");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_F32) {
@@ -152,8 +158,9 @@ TEST_F(MslGeneratorImplTest, EmitType_F32) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(f32, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "float");
+  ASSERT_TRUE(gen.EmitType(out, f32, "")) << gen.error();
+  EXPECT_EQ(out.str(), "float");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_I32) {
@@ -161,8 +168,9 @@ TEST_F(MslGeneratorImplTest, EmitType_I32) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(i32, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "int");
+  ASSERT_TRUE(gen.EmitType(out, i32, "")) << gen.error();
+  EXPECT_EQ(out.str(), "int");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Matrix) {
@@ -172,8 +180,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Matrix) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(mat2x3, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "float2x3");
+  ASSERT_TRUE(gen.EmitType(out, mat2x3, "")) << gen.error();
+  EXPECT_EQ(out.str(), "float2x3");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Pointer) {
@@ -183,8 +192,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Pointer) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(p, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "threadgroup float* ");
+  ASSERT_TRUE(gen.EmitType(out, p, "")) << gen.error();
+  EXPECT_EQ(out.str(), "threadgroup float* ");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Struct) {
@@ -195,8 +205,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Struct) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(s), "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "S");
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(s), "")) << gen.error();
+  EXPECT_EQ(out.str(), "S");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_StructDecl) {
@@ -670,8 +681,9 @@ TEST_F(MslGeneratorImplTest, DISABLED_EmitType_Struct_WithDecoration) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(s), "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), R"(struct {
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(s), "")) << gen.error();
+  EXPECT_EQ(out.str(), R"(struct {
   /* 0x0000 */ int a;
   /* 0x0004 */ float b;
 })");
@@ -682,8 +694,9 @@ TEST_F(MslGeneratorImplTest, EmitType_U32) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(u32, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "uint");
+  ASSERT_TRUE(gen.EmitType(out, u32, "")) << gen.error();
+  EXPECT_EQ(out.str(), "uint");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Vector) {
@@ -692,8 +705,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Vector) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(vec3, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "float3");
+  ASSERT_TRUE(gen.EmitType(out, vec3, "")) << gen.error();
+  EXPECT_EQ(out.str(), "float3");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Void) {
@@ -701,8 +715,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Void) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(void_, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "void");
+  ASSERT_TRUE(gen.EmitType(out, void_, "")) << gen.error();
+  EXPECT_EQ(out.str(), "void");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_Sampler) {
@@ -710,8 +725,9 @@ TEST_F(MslGeneratorImplTest, EmitType_Sampler) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(sampler, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "sampler");
+  ASSERT_TRUE(gen.EmitType(out, sampler, "")) << gen.error();
+  EXPECT_EQ(out.str(), "sampler");
 }
 
 TEST_F(MslGeneratorImplTest, EmitType_SamplerComparison) {
@@ -719,8 +735,9 @@ TEST_F(MslGeneratorImplTest, EmitType_SamplerComparison) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(sampler, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "sampler");
+  ASSERT_TRUE(gen.EmitType(out, sampler, "")) << gen.error();
+  EXPECT_EQ(out.str(), "sampler");
 }
 
 struct MslDepthTextureData {
@@ -739,8 +756,9 @@ TEST_P(MslDepthTexturesTest, Emit) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(&s, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), params.result);
+  ASSERT_TRUE(gen.EmitType(out, &s, "")) << gen.error();
+  EXPECT_EQ(out.str(), params.result);
 }
 INSTANTIATE_TEST_SUITE_P(
     MslGeneratorImplTest,
@@ -772,8 +790,9 @@ TEST_P(MslSampledtexturesTest, Emit) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(s, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), params.result);
+  ASSERT_TRUE(gen.EmitType(out, s, "")) << gen.error();
+  EXPECT_EQ(out.str(), params.result);
 }
 INSTANTIATE_TEST_SUITE_P(
     MslGeneratorImplTest,
@@ -798,8 +817,9 @@ TEST_F(MslGeneratorImplTest, Emit_TypeMultisampledTexture) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(ms, "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "texture2d_ms<uint, access::read>");
+  ASSERT_TRUE(gen.EmitType(out, ms, "")) << gen.error();
+  EXPECT_EQ(out.str(), "texture2d_ms<uint, access::read>");
 }
 
 struct MslStorageTextureData {
@@ -826,8 +846,9 @@ TEST_P(MslStorageTexturesTest, Emit) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitType(program->TypeOf(s), "")) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), params.result);
+  ASSERT_TRUE(gen.EmitType(out, program->TypeOf(s), "")) << gen.error();
+  EXPECT_EQ(out.str(), params.result);
 }
 INSTANTIATE_TEST_SUITE_P(
     MslGeneratorImplTest,

src/writer/msl/generator_impl_unary_op_test.cc

@@ -29,8 +29,9 @@ TEST_F(MslUnaryOpTest, AddressOf) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(op)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "&(expr)");
+  ASSERT_TRUE(gen.EmitExpression(out, op)) << gen.error();
+  EXPECT_EQ(out.str(), "&(expr)");
 }
 
 TEST_F(MslUnaryOpTest, Complement) {
@@ -41,8 +42,9 @@ TEST_F(MslUnaryOpTest, Complement) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(op)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "~(expr)");
+  ASSERT_TRUE(gen.EmitExpression(out, op)) << gen.error();
+  EXPECT_EQ(out.str(), "~(expr)");
 }
 
 TEST_F(MslUnaryOpTest, Indirection) {
@@ -56,8 +58,9 @@ TEST_F(MslUnaryOpTest, Indirection) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(op)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "*(expr)");
+  ASSERT_TRUE(gen.EmitExpression(out, op)) << gen.error();
+  EXPECT_EQ(out.str(), "*(expr)");
 }
 
 TEST_F(MslUnaryOpTest, Not) {
@@ -67,8 +70,9 @@ TEST_F(MslUnaryOpTest, Not) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(op)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "!(expr)");
+  ASSERT_TRUE(gen.EmitExpression(out, op)) << gen.error();
+  EXPECT_EQ(out.str(), "!(expr)");
 }
 
 TEST_F(MslUnaryOpTest, Negation) {
@@ -79,8 +83,9 @@ TEST_F(MslUnaryOpTest, Negation) {
 
   GeneratorImpl& gen = Build();
 
-  ASSERT_TRUE(gen.EmitExpression(op)) << gen.error();
+  std::stringstream out;
-  EXPECT_EQ(gen.result(), "-(expr)");
+  ASSERT_TRUE(gen.EmitExpression(out, op)) << gen.error();
+  EXPECT_EQ(out.str(), "-(expr)");
 }
 
 }  // namespace

src/writer/text_generator.cc

@@ -44,8 +44,9 @@ TextGenerator::LineWriter::~LineWriter() {
     auto str = os.str();
     if (!str.empty()) {
       gen->make_indent();
-      gen->out_ << str << std::endl;
+      gen->out_ << str;
     }
+    gen->out_ << std::endl;
   }
 }
 

test/bug/dawn/947.wgsl.expected.msl

@@ -39,8 +39,8 @@ vertex tint_symbol_1 vs_main(uint VertexIndex [[vertex_id]], constant Uniforms&
 
 fragment tint_symbol_4 fs_main(texture2d<float, access::sample> tint_symbol_7 [[texture(2)]], sampler tint_symbol_8 [[sampler(1)]], tint_symbol_3 tint_symbol_2 [[stage_in]]) {
   float2 const texcoord = tint_symbol_2.texcoord;
-  float2 clampedTexcoord =   clamp(texcoord, float2(0.0f, 0.0f), float2(1.0f, 1.0f));
+  float2 clampedTexcoord = clamp(texcoord, float2(0.0f, 0.0f), float2(1.0f, 1.0f));
-  if (!(  all((clampedTexcoord == texcoord)))) {
+  if (!(all((clampedTexcoord == texcoord)))) {
     discard_fragment();
   }
   float4 srcColor = tint_symbol_7.sample(tint_symbol_8, texcoord);

test/bug/tint/749.spvasm.expected.msl

@@ -1451,7 +1451,7 @@ void main_1(constant buf0& x_188, thread QuicksortObject* const tint_symbol_86,
   int const x_909 = (*(tint_symbol_86)).numbers.arr[4];
   (*(tint_symbol_86)).numbers.arr[4] = 0;
   (*(tint_symbol_86)).numbers.arr[4] = x_909;
-  if ((  fabs((x_308 - x_310)) < 0.25f)) {
+  if ((fabs((x_308 - x_310)) < 0.25f)) {
     float const x_910 = uv.x;
     uv.x = 0.0f;
     uv.x = x_910;
@@ -1505,7 +1505,7 @@ void main_1(constant buf0& x_188, thread QuicksortObject* const tint_symbol_86,
   uv[0] = 0.0f;
   uv[0] = x_922;
   float3 const x_519 = float3(x_447.x, x_446.x, x_446.y);
-  float3 const x_326 =   normalize(x_325);
+  float3 const x_326 = normalize(x_325);
   float const x_923 = uv.x;
   uv.x = 0.0f;
   uv.x = x_923;

test/bug/tint/913.wgsl.expected.msl

@@ -13,7 +13,7 @@ struct OutputBuf {
 };
 
 bool aboutEqual(float value, float expect) {
-  return (  fabs((value - expect)) < 0.001f);
+  return (fabs((value - expect)) < 0.001f);
 }
 
 kernel void tint_symbol(texture2d<float, access::sample> tint_symbol_2 [[texture(0)]], texture2d<float, access::sample> tint_symbol_3 [[texture(1)]], uint3 GlobalInvocationID [[thread_position_in_grid]], constant Uniforms& uniforms [[buffer(3)]], device OutputBuf& output [[buffer(2)]]) {
@@ -23,7 +23,7 @@ kernel void tint_symbol(texture2d<float, access::sample> tint_symbol_2 [[texture
   float4 const nonCoveredColor = float4(0.0f, 1.0f, 0.0f, 1.0f);
   bool success = true;
   if (((((dstTexCoord.x < uniforms.dstCopyOrigin.x) || (dstTexCoord.y < uniforms.dstCopyOrigin.y)) || (dstTexCoord.x >= (uniforms.dstCopyOrigin.x + uniforms.copySize.x))) || (dstTexCoord.y >= (uniforms.dstCopyOrigin.y + uniforms.copySize.y)))) {
-    success = (success &&     all((tint_symbol_3.read(uint2(int2(dstTexCoord)), 0) == nonCoveredColor)));
+    success = (success && all((tint_symbol_3.read(uint2(int2(dstTexCoord)), 0) == nonCoveredColor)));
   } else {
     uint2 srcTexCoord = ((dstTexCoord - uniforms.dstCopyOrigin) + uniforms.srcCopyOrigin);
     if ((uniforms.dstTextureFlipY == 1u)) {

test/bug/tint/914.wgsl.expected.msl

@@ -59,7 +59,7 @@ kernel void tint_symbol(uint3 local_id [[thread_position_in_threadgroup]], uint3
     tint_array_wrapper const tint_symbol_3 = {.arr={}};
     tint_symbol_5 = tint_symbol_3;
   }
-    threadgroup_barrier(mem_flags::mem_threadgroup);
+  threadgroup_barrier(mem_flags::mem_threadgroup);
   uint const tileRow = (local_id.y * RowPerThread);
   uint const tileCol = (local_id.x * ColPerThread);
   uint const globalRow = (global_id.y * RowPerThread);
@@ -140,7 +140,7 @@ kernel void tint_symbol(uint3 local_id [[thread_position_in_threadgroup]], uint3
           }
         }
       }
-            threadgroup_barrier(mem_flags::mem_threadgroup);
+      threadgroup_barrier(mem_flags::mem_threadgroup);
       {
         uint k = 0u;
         while (true) {
@@ -189,7 +189,7 @@ kernel void tint_symbol(uint3 local_id [[thread_position_in_threadgroup]], uint3
           }
         }
       }
-            threadgroup_barrier(mem_flags::mem_threadgroup);
+      threadgroup_barrier(mem_flags::mem_threadgroup);
       {
         t = (t + 1u);
       }

test/intrinsics/gen/abs/002533.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_002533() {
-  float4 res =   fabs(float4());
+  float4 res = fabs(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/005174.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_005174() {
-  float3 res =   fabs(float3());
+  float3 res = fabs(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/1ce782.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_1ce782() {
-  uint4 res =   abs(uint4());
+  uint4 res = abs(uint4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/1e9d53.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_1e9d53() {
-  float2 res =   fabs(float2());
+  float2 res = fabs(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/467cd1.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_467cd1() {
-  uint res =   abs(1u);
+  uint res = abs(1u);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/4ad288.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_4ad288() {
-  int res =   abs(1);
+  int res = abs(1);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/5ad50a.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_5ad50a() {
-  int3 res =   abs(int3());
+  int3 res = abs(int3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/7326de.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_7326de() {
-  uint3 res =   abs(uint3());
+  uint3 res = abs(uint3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/7f28e6.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_7f28e6() {
-  uint2 res =   abs(uint2());
+  uint2 res = abs(uint2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/7faa9e.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_7faa9e() {
-  int2 res =   abs(int2());
+  int2 res = abs(int2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/9c80a6.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_9c80a6() {
-  int4 res =   abs(int4());
+  int4 res = abs(int4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/abs/b96037.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void abs_b96037() {
-  float res =   fabs(1.0f);
+  float res = fabs(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/acos/489247.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void acos_489247() {
-  float res =   acos(1.0f);
+  float res = acos(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/acos/8e2acf.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void acos_8e2acf() {
-  float4 res =   acos(float4());
+  float4 res = acos(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/acos/a610c4.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void acos_a610c4() {
-  float3 res =   acos(float3());
+  float3 res = acos(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/acos/dfc915.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void acos_dfc915() {
-  float2 res =   acos(float2());
+  float2 res = acos(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/all/986c7b.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void all_986c7b() {
-  bool res =   all(bool4());
+  bool res = all(bool4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/all/bd2dba.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void all_bd2dba() {
-  bool res =   all(bool3());
+  bool res = all(bool3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/all/f46790.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void all_f46790() {
-  bool res =   all(bool2());
+  bool res = all(bool2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/any/083428.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void any_083428() {
-  bool res =   any(bool4());
+  bool res = any(bool4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/any/0e3e58.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void any_0e3e58() {
-  bool res =   any(bool2());
+  bool res = any(bool2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/any/e755c1.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void any_e755c1() {
-  bool res =   any(bool3());
+  bool res = any(bool3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/asin/064953.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void asin_064953() {
-  float4 res =   asin(float4());
+  float4 res = asin(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/asin/7b6a44.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void asin_7b6a44() {
-  float2 res =   asin(float2());
+  float2 res = asin(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/asin/8cd9c9.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void asin_8cd9c9() {
-  float3 res =   asin(float3());
+  float3 res = asin(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/asin/c0c272.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void asin_c0c272() {
-  float res =   asin(1.0f);
+  float res = asin(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan/02979a.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan_02979a() {
-  float res =   atan(1.0f);
+  float res = atan(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan/331e6d.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan_331e6d() {
-  float3 res =   atan(float3());
+  float3 res = atan(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan/a8b696.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan_a8b696() {
-  float4 res =   atan(float4());
+  float4 res = atan(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan/ad96e4.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan_ad96e4() {
-  float2 res =   atan(float2());
+  float2 res = atan(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan2/57fb13.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan2_57fb13() {
-  float2 res =   atan2(float2(), float2());
+  float2 res = atan2(float2(), float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan2/96057c.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan2_96057c() {
-  float res =   atan2(1.0f, 1.0f);
+  float res = atan2(1.0f, 1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan2/a70d0d.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan2_a70d0d() {
-  float3 res =   atan2(float3(), float3());
+  float3 res = atan2(float3(), float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/atan2/ae713e.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void atan2_ae713e() {
-  float4 res =   atan2(float4(), float4());
+  float4 res = atan2(float4(), float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/ceil/34064b.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void ceil_34064b() {
-  float3 res =   ceil(float3());
+  float3 res = ceil(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/ceil/678655.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void ceil_678655() {
-  float res =   ceil(1.0f);
+  float res = ceil(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/ceil/96f597.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void ceil_96f597() {
-  float2 res =   ceil(float2());
+  float2 res = ceil(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/ceil/b74c16.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void ceil_b74c16() {
-  float4 res =   ceil(float4());
+  float4 res = ceil(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/0acf8f.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_0acf8f() {
-  float2 res =   clamp(float2(), float2(), float2());
+  float2 res = clamp(float2(), float2(), float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/1a32e3.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_1a32e3() {
-  int4 res =   clamp(int4(), int4(), int4());
+  int4 res = clamp(int4(), int4(), int4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/2bd567.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_2bd567() {
-  float res =   clamp(1.0f, 1.0f, 1.0f);
+  float res = clamp(1.0f, 1.0f, 1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/2bde41.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_2bde41() {
-  float4 res =   clamp(float4(), float4(), float4());
+  float4 res = clamp(float4(), float4(), float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/548fc7.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_548fc7() {
-  uint3 res =   clamp(uint3(), uint3(), uint3());
+  uint3 res = clamp(uint3(), uint3(), uint3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/5f0819.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_5f0819() {
-  int3 res =   clamp(int3(), int3(), int3());
+  int3 res = clamp(int3(), int3(), int3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/6c1749.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_6c1749() {
-  int2 res =   clamp(int2(), int2(), int2());
+  int2 res = clamp(int2(), int2(), int2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/7706d7.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_7706d7() {
-  uint2 res =   clamp(uint2(), uint2(), uint2());
+  uint2 res = clamp(uint2(), uint2(), uint2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/867397.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_867397() {
-  float3 res =   clamp(float3(), float3(), float3());
+  float3 res = clamp(float3(), float3(), float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/a2de25.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_a2de25() {
-  uint res =   clamp(1u, 1u, 1u);
+  uint res = clamp(1u, 1u, 1u);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/b07c65.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_b07c65() {
-  int res =   clamp(1, 1, 1);
+  int res = clamp(1, 1, 1);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/clamp/bd43ce.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void clamp_bd43ce() {
-  uint4 res =   clamp(uint4(), uint4(), uint4());
+  uint4 res = clamp(uint4(), uint4(), uint4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cos/16dc15.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cos_16dc15() {
-  float3 res =   cos(float3());
+  float3 res = cos(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cos/29d66d.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cos_29d66d() {
-  float4 res =   cos(float4());
+  float4 res = cos(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cos/c3b486.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cos_c3b486() {
-  float2 res =   cos(float2());
+  float2 res = cos(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cos/c5c28e.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cos_c5c28e() {
-  float res =   cos(1.0f);
+  float res = cos(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cosh/377652.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cosh_377652() {
-  float3 res =   cosh(float3());
+  float3 res = cosh(float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cosh/c13756.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cosh_c13756() {
-  float2 res =   cosh(float2());
+  float2 res = cosh(float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cosh/da92dd.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cosh_da92dd() {
-  float res =   cosh(1.0f);
+  float res = cosh(1.0f);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cosh/e0c1de.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cosh_e0c1de() {
-  float4 res =   cosh(float4());
+  float4 res = cosh(float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/0d0e46.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_0d0e46() {
-  uint4 res =   popcount(uint4());
+  uint4 res = popcount(uint4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/0f7980.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_0f7980() {
-  int4 res =   popcount(int4());
+  int4 res = popcount(int4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/65d2ae.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_65d2ae() {
-  int3 res =   popcount(int3());
+  int3 res = popcount(int3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/690cfc.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_690cfc() {
-  uint3 res =   popcount(uint3());
+  uint3 res = popcount(uint3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/94fd81.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_94fd81() {
-  uint2 res =   popcount(uint2());
+  uint2 res = popcount(uint2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/ae44f9.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_ae44f9() {
-  uint res =   popcount(1u);
+  uint res = popcount(1u);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/af90e2.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_af90e2() {
-  int2 res =   popcount(int2());
+  int2 res = popcount(int2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/countOneBits/fd88b2.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void countOneBits_fd88b2() {
-  int res =   popcount(1);
+  int res = popcount(1);
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/cross/041cb0.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void cross_041cb0() {
-  float3 res =   cross(float3(), float3());
+  float3 res = cross(float3(), float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/determinant/2b62ba.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void determinant_2b62ba() {
-  float res =   determinant(float3x3());
+  float res = determinant(float3x3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/determinant/a0a87c.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void determinant_a0a87c() {
-  float res =   determinant(float4x4());
+  float res = determinant(float4x4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/determinant/e19305.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void determinant_e19305() {
-  float res =   determinant(float2x2());
+  float res = determinant(float2x2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/distance/0657d4.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void distance_0657d4() {
-  float res =   distance(float3(), float3());
+  float res = distance(float3(), float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/distance/9646ea.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void distance_9646ea() {
-  float res =   distance(float4(), float4());
+  float res = distance(float4(), float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/distance/aa4055.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void distance_aa4055() {
-  float res =   distance(float2(), float2());
+  float res = distance(float2(), float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/dot/0c577b.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void dot_0c577b() {
-  float res =   dot(float4(), float4());
+  float res = dot(float4(), float4());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/dot/883f0e.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void dot_883f0e() {
-  float res =   dot(float2(), float2());
+  float res = dot(float2(), float2());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/dot/ba4246.wgsl.expected.msl

@@ -6,7 +6,7 @@ struct tint_symbol {
 };
 
 void dot_ba4246() {
-  float res =   dot(float3(), float3());
+  float res = dot(float3(), float3());
 }
 
 vertex tint_symbol vertex_main() {

test/intrinsics/gen/dpdx/0763f7.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void dpdx_0763f7() {
-  float3 res =   dfdx(float3());
+  float3 res = dfdx(float3());
 }
 
 fragment void fragment_main() {

test/intrinsics/gen/dpdx/99edb1.wgsl.expected.msl

@@ -2,7 +2,7 @@
 
 using namespace metal;
 void dpdx_99edb1() {
-  float2 res =   dfdx(float2());
+  float2 res = dfdx(float2());
 }
 
 fragment void fragment_main() {