[spirv-reader][ir] Support OpCopyObject

Handle the SPIR-V `OpCopyObject` instruction. This just treats the
`OpCopyObject` result id as the same as the copied `Value`.

Bug: 42250952
Change-Id: I3e98a28dd4c656adf906517a82996444118f9655
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/220497
Commit-Queue: dan sinclair <dsinclair@chromium.org>
Reviewed-by: James Price <jrprice@google.com>

src/tint/lang/core/ir/builder.h

diff --git a/src/tint/lang/core/ir/builder.h b/src/tint/lang/core/ir/builder.h
index f411d64..59f6050 100644
--- a/src/tint/lang/core/ir/builder.h
+++ b/src/tint/lang/core/ir/builder.h
@@ -1426,6 +1426,24 @@
         return let;
     }
 
+    /// Creates a new `let` declaration
+    /// @param value the value
+    /// @returns the instruction
+    template <
+        typename VALUE,
+        typename = std::enable_if_t<
+            !traits::IsTypeOrDerived<std::remove_pointer_t<std::decay_t<VALUE>>, core::type::Type>>>
+    ir::Let* Let(VALUE&& value) {
+        auto* val = Value(std::forward<VALUE>(value));
+        if (DAWN_UNLIKELY(!val)) {
+            TINT_ASSERT(val);
+            return nullptr;
+        }
+        auto* let = ir.CreateInstruction<ir::Let>(InstructionResult(val->Type()), val);
+        Append(let);
+        return let;
+    }
+
     /// Creates a return instruction
     /// @param func the function being returned
     /// @returns the instruction

src/tint/lang/spirv/reader/parser/misc_test.cc

diff --git a/src/tint/lang/spirv/reader/parser/misc_test.cc b/src/tint/lang/spirv/reader/parser/misc_test.cc
index 6ef1a61..96ea15ab 100644
--- a/src/tint/lang/spirv/reader/parser/misc_test.cc
+++ b/src/tint/lang/spirv/reader/parser/misc_test.cc
@@ -114,5 +114,75 @@
 )");
 }
 
+TEST_F(SpirvParserTest, Misc_OpCopyObject) {
+    EXPECT_IR(R"(
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint GLCompute %main "main"
+               OpExecutionMode %main LocalSize 1 1 1
+       %void = OpTypeVoid
+       %bool = OpTypeBool
+       %uint = OpTypeInt 32 0
+        %int = OpTypeInt 32 1
+      %float = OpTypeFloat 32
+     %uint_2 = OpConstant %uint 2
+%_arr_uint_uint_2 = OpTypeArray %uint %uint_2
+     %v2uint = OpTypeVector %uint 2
+      %v2int = OpTypeVector %int 2
+    %v2float = OpTypeVector %float 2
+%mat2v2float = OpTypeMatrix %v2float 2
+ %_struct_13 = OpTypeStruct %bool %uint %int %float
+    %ep_type = OpTypeFunction %void
+       %main = OpFunction %void None %ep_type
+ %main_start = OpLabel
+          %1 = OpUndef %bool
+          %2 = OpUndef %uint
+          %3 = OpUndef %int
+          %4 = OpUndef %float
+          %5 = OpUndef %_arr_uint_uint_2
+          %6 = OpUndef %mat2v2float
+          %7 = OpUndef %v2uint
+          %8 = OpUndef %v2int
+          %9 = OpUndef %v2float
+         %10 = OpUndef %_struct_13
+         %11 = OpCopyObject %bool %1
+         %12 = OpCopyObject %uint %2
+         %13 = OpCopyObject %int %3
+         %14 = OpCopyObject %float %4
+         %15 = OpCopyObject %_arr_uint_uint_2 %5
+         %16 = OpCopyObject %mat2v2float %6
+         %17 = OpCopyObject %v2uint %7
+         %18 = OpCopyObject %v2int %8
+         %19 = OpCopyObject %v2float %9
+         %20 = OpCopyObject %_struct_13 %10
+               OpReturn
+               OpFunctionEnd
+)",
+              R"(
+tint_symbol_4 = struct @align(4) {
+  tint_symbol:bool @offset(0)
+  tint_symbol_1:u32 @offset(4)
+  tint_symbol_2:i32 @offset(8)
+  tint_symbol_3:f32 @offset(12)
+}
+
+%main = @compute @workgroup_size(1u, 1u, 1u) func():void {
+  $B1: {
+    %2:bool = let false
+    %3:u32 = let 0u
+    %4:i32 = let 0i
+    %5:f32 = let 0.0f
+    %6:array<u32, 2> = let array<u32, 2>(0u)
+    %7:mat2x2<f32> = let mat2x2<f32>(vec2<f32>(0.0f))
+    %8:vec2<u32> = let vec2<u32>(0u)
+    %9:vec2<i32> = let vec2<i32>(0i)
+    %10:vec2<f32> = let vec2<f32>(0.0f)
+    %11:tint_symbol_4 = let tint_symbol_4(false, 0u, 0i, 0.0f)
+    ret
+  }
+}
+)");
+}
+
 }  // namespace
 }  // namespace tint::spirv::reader

src/tint/lang/spirv/reader/parser/parser.cc

diff --git a/src/tint/lang/spirv/reader/parser/parser.cc b/src/tint/lang/spirv/reader/parser/parser.cc
index 418f7fe..2b20fbb 100644
--- a/src/tint/lang/spirv/reader/parser/parser.cc
+++ b/src/tint/lang/spirv/reader/parser/parser.cc
@@ -507,6 +507,9 @@
                 case spv::Op::OpUndef:
                     AddValue(inst.result_id(), b_.Zero(Type(inst.type_id())));
                     break;
+                case spv::Op::OpCopyObject:
+                    EmitCopyObject(inst);
+                    break;
                 case spv::Op::OpAccessChain:
                 case spv::Op::OpInBoundsAccessChain:
                     EmitAccess(inst);
@@ -552,6 +555,13 @@
         }
     }
 
+    /// @param inst the SPIR-V instruction for OpCopyObject
+    void EmitCopyObject(const spvtools::opt::Instruction& inst) {
+        // Make the result Id a pointer to the original copied value.
+        auto* l = b_.Let(Value(inst.GetSingleWordOperand(2)));
+        Emit(l, inst.result_id());
+    }
+
     /// @param inst the SPIR-V instruction for OpAccessChain
     void EmitAccess(const spvtools::opt::Instruction& inst) {
         Vector<core::ir::Value*, 4> indices;