[ir] Fix scope stack for loops.

The scope stack for a loop is slightly different from other constructs
as the block for the body and the block for the continuing need to share
the scope. This CL fixes the loop conversion to special case the scope
stack for the loop body so the continuing block can see the variables.

Bug: tint:1718
Change-Id: I2e4898b36f1541b48a4e349955833b155332947a
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/133160
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Dan Sinclair <dsinclair@chromium.org>

src/tint/ir/from_program.cc

@@ -409,7 +409,6 @@ class Impl {
         if (!rhs) {
             return;
         }
-
         auto* ty = lhs.Get()->Type();
         Binary* inst = nullptr;
         switch (stmt->op) {
@@ -523,7 +522,12 @@ class Impl {
             FlowStackScope scope(this, loop_node);
 
             current_flow_block_ = loop_node->start.target->As<Block>();
-            EmitBlock(stmt->body);
+
+            // The loop doesn't use EmitBlock because it needs the scope stack to not get popped
+            // until after the continuing block.
+            scopes_.Push();
+            TINT_DEFER(scopes_.Pop());
+            EmitStatements(stmt->body->statements);
 
             // The current block didn't `break`, `return` or `continue`, go to the continuing block.
             BranchToIfNeeded(loop_node->continuing.target);
@@ -782,9 +786,14 @@ class Impl {
             [&](const ast::BinaryExpression* b) { return EmitBinary(b); },
             [&](const ast::BitcastExpression* b) { return EmitBitcast(b); },
             [&](const ast::CallExpression* c) { return EmitCall(c); },
-            [&](const ast::IdentifierExpression* i) {
+            [&](const ast::IdentifierExpression* i) -> utils::Result<Value*> {
                 auto* v = scopes_.Get(i->identifier->symbol);
-                return utils::Result<Value*>{v};
+                if (TINT_UNLIKELY(!v)) {
+                    add_error(expr->source,
+                              "unable to find identifier " + i->identifier->symbol.Name());
+                    return utils::Failure;
+                }
+                return {v};
             },
             [&](const ast::LiteralExpression* l) { return EmitLiteral(l); },
             // [&](const ast::MemberAccessorExpression* m) {

src/tint/ir/from_program_test.cc

@@ -482,6 +482,50 @@ func_end
 )");
 }
 
+TEST_F(IR_BuilderImplTest, Loop_Continuing_Body_Scope) {
+    auto* a = Decl(Let("a", Expr(true)));
+    auto* ast_break_if = BreakIf("a");
+    auto* ast_loop = Loop(Block(a), Block(ast_break_if));
+    WrapInFunction(ast_loop);
+
+    auto m = Build();
+    ASSERT_TRUE(m) << (!m ? m.Failure() : "");
+
+    EXPECT_EQ(Disassemble(m.Get()),
+              R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
+  %fn2 = block
+  branch %fn3
+
+  %fn3 = loop [s: %fn4, c: %fn5, m: %fn6]
+    # loop start
+    %fn4 = block
+    branch %fn5
+
+    # loop continuing
+    %fn5 = block
+    branch %fn7
+
+    %fn7 = if true [t: %fn8, f: %fn9, m: %fn10]
+      # true branch
+      %fn8 = block
+      branch %fn6
+
+      # false branch
+      %fn9 = block
+      branch %fn10
+
+    # if merge
+    %fn10 = block
+    branch %fn4
+
+  # loop merge
+  %fn6 = block
+  ret
+func_end
+
+)");
+}
+
 TEST_F(IR_BuilderImplTest, Loop_WithReturn) {
     auto* ast_if = If(true, Block(Return()));
     auto* ast_loop = Loop(Block(ast_if, Continue()));