main: Replace --dawn-validation with --validate

Performs output validation with spirv-val for SPIR-V (as before), HLSL
validation with DXC, and MSL validation with the Metal Shader Compiler.

Disable HLSL tests that fail to validate

Bug: tint:812
Change-Id: If78c351b4e23c7fb50d333eacf9ee7cc81d18564
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/51280
Reviewed-by: Antonio Maiorano <amaiorano@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Antonio Maiorano <amaiorano@google.com>

samples/BUILD.gn

@@ -19,6 +19,7 @@ executable("tint") {
   sources = [ "main.cc" ]
   deps = [
     "${tint_root_dir}/src:libtint",
+    "${tint_root_dir}/src:tint_val",
     "${tint_spirv_tools_dir}/:spvtools",
     "${tint_spirv_tools_dir}/:spvtools_opt",
     "${tint_spirv_tools_dir}/:spvtools_val",

samples/CMakeLists.txt

@@ -19,7 +19,7 @@ set(TINT_SRCS
 ## Tint executable
 add_executable(tint ${TINT_SRCS})
 tint_default_compile_options(tint)
-target_link_libraries(tint libtint)
+target_link_libraries(tint libtint tint_val)
 
 if(${TINT_BUILD_SPV_READER} OR ${TINT_BUILD_SPV_WRITER})
   target_link_libraries(tint SPIRV-Tools)

samples/main.cc

@@ -24,6 +24,8 @@
 #include "spirv-tools/libspirv.hpp"
 #endif  // TINT_BUILD_SPV_READER
 
+#include "src/utils/io/command.h"
+#include "src/val/val.h"
 #include "tint/tint.h"
 
 namespace {
@@ -52,7 +54,7 @@ struct Options {
 
   bool parse_only = false;
   bool dump_ast = false;
-  bool dawn_validation = false;
+  bool validate = false;
   bool demangle = false;
   bool dump_inspector_bindings = false;
 
@@ -62,6 +64,9 @@ struct Options {
   std::string ep_name;
 
   std::vector<std::string> transforms;
+
+  std::string dxc_path;
+  std::string xcrun_path;
 };
 
 const char kUsage[] = R"(Usage: tint [options] <input-file>
@@ -86,12 +91,16 @@ const char kUsage[] = R"(Usage: tint [options] <input-file>
                                 renamer
   --parse-only              -- Stop after parsing the input
   --dump-ast                -- Dump the generated AST to stdout
-  --dawn-validation         -- SPIRV outputs are validated with the same flags
-                               as Dawn does. Has no effect on non-SPIRV outputs.
   --demangle                -- Preserve original source names. Demangle them.
                                Affects AST dumping, and text-based output languages.
   --dump-inspector-bindings -- Dump reflection data about bindins to stdout.
-  -h                        -- This help text)";
+  -h                        -- This help text
+  --validate                -- Validates generated SPIR-V with spirv-val.
+                               Has no effect on non-SPIRV outputs.
+  --dxc                     -- Path to DXC executable, used to validate HLSL output.
+                               When specified, automatically enables --validate
+  --xcrun                   -- Path to xcrun executable, used to validate MSL output.
+                               When specified, automatically enables --validate)";
 
 #ifdef _MSC_VER
 #pragma warning(disable : 4068; suppress : 4100)
@@ -389,12 +398,28 @@ bool ParseArgs(const std::vector<std::string>& args, Options* opts) {
       opts->parse_only = true;
     } else if (arg == "--dump-ast") {
       opts->dump_ast = true;
-    } else if (arg == "--dawn-validation") {
-      opts->dawn_validation = true;
     } else if (arg == "--demangle") {
       opts->demangle = true;
     } else if (arg == "--dump-inspector-bindings") {
       opts->dump_inspector_bindings = true;
+    } else if (arg == "--validate") {
+      opts->validate = true;
+    } else if (arg == "--dxc") {
+      ++i;
+      if (i >= args.size()) {
+        std::cerr << "Missing value for " << arg << std::endl;
+        return false;
+      }
+      opts->dxc_path = args[i];
+      opts->validate = true;
+    } else if (arg == "--xcrun") {
+      ++i;
+      if (i >= args.size()) {
+        std::cerr << "Missing value for " << arg << std::endl;
+        return false;
+      }
+      opts->xcrun_path = args[i];
+      opts->validate = true;
     } else if (!arg.empty()) {
       if (arg[0] == '-') {
         std::cerr << "Unrecognized option: " << arg << std::endl;
@@ -772,31 +797,45 @@ int main(int argc, const char** argv) {
 
   std::unique_ptr<tint::writer::Writer> writer;
 
+  switch (options.format) {
+    case Format::kSpirv:
+    case Format::kSpvAsm:
 #if TINT_BUILD_SPV_WRITER
-  if (options.format == Format::kSpirv || options.format == Format::kSpvAsm) {
       writer = std::make_unique<tint::writer::spirv::Generator>(program.get());
-  }
+      break;
+#else
+      std::cerr << "SPIR-V writer not enabled in tint build" << std::endl;
+      return 1;
 #endif  // TINT_BUILD_SPV_WRITER
 
+    case Format::kWgsl:
 #if TINT_BUILD_WGSL_WRITER
-  if (options.format == Format::kWgsl) {
       writer = std::make_unique<tint::writer::wgsl::Generator>(program.get());
-  }
+      break;
+#else
+      std::cerr << "WGSL writer not enabled in tint build" << std::endl;
+      return 1;
 #endif  // TINT_BUILD_WGSL_WRITER
 
+    case Format::kMsl:
 #if TINT_BUILD_MSL_WRITER
-  if (options.format == Format::kMsl) {
       writer = std::make_unique<tint::writer::msl::Generator>(program.get());
-  }
+      break;
+#else
+      std::cerr << "MSL writer not enabled in tint build" << std::endl;
+      return 1;
 #endif  // TINT_BUILD_MSL_WRITER
 
+    case Format::kHlsl:
 #if TINT_BUILD_HLSL_WRITER
-  if (options.format == Format::kHlsl) {
       writer = std::make_unique<tint::writer::hlsl::Generator>(program.get());
-  }
+      break;
+#else
+      std::cerr << "HLSL writer not enabled in tint build" << std::endl;
+      return 1;
 #endif  // TINT_BUILD_HLSL_WRITER
 
-  if (!writer) {
+    default:
       std::cerr << "Unknown output format specified" << std::endl;
       return 1;
   }
@@ -807,27 +846,82 @@ int main(int argc, const char** argv) {
     return 1;
   }
 
-#if TINT_BUILD_SPV_WRITER
+  bool validation_failed = false;
-  bool dawn_validation_failed = false;
+  std::ostringstream validation_msgs;
-  std::ostringstream stream;
 
-  if (options.dawn_validation &&
+  if (options.validate) {
-      (options.format == Format::kSpvAsm || options.format == Format::kSpirv)) {
+    switch (options.format) {
+#if TINT_BUILD_SPV_WRITER
+      case Format::kSpirv:
+      case Format::kSpvAsm: {
         // Use Vulkan 1.1, since this is what Tint, internally, uses.
         spvtools::SpirvTools tools(SPV_ENV_VULKAN_1_1);
-    tools.SetMessageConsumer([&stream](spv_message_level_t, const char*,
+        tools.SetMessageConsumer(
-                                       const spv_position_t& pos,
+            [&validation_msgs](spv_message_level_t, const char*,
-                                       const char* msg) {
+                               const spv_position_t& pos, const char* msg) {
-      stream << (pos.line + 1) << ":" << (pos.column + 1) << ": " << msg
+              validation_msgs << (pos.line + 1) << ":" << (pos.column + 1)
-             << std::endl;
+                              << ": " << msg << std::endl;
             });
         auto* w = static_cast<tint::writer::spirv::Generator*>(writer.get());
         if (!tools.Validate(w->result().data(), w->result().size(),
                             spvtools::ValidatorOptions())) {
-      dawn_validation_failed = true;
+          validation_failed = true;
+        }
+        break;
+      }
+#endif
+#if TINT_BUILD_HLSL_WRITER
+      case Format::kHlsl: {
+        auto dxc = tint::utils::Command::LookPath(
+            options.dxc_path.empty() ? "dxc" : options.dxc_path);
+        if (dxc.Found()) {
+          auto* w = static_cast<tint::writer::Text*>(writer.get());
+          auto hlsl = w->result();
+          auto res = tint::val::Hlsl(dxc.Path(), hlsl, program.get());
+          if (res.failed) {
+            validation_failed = true;
+            validation_msgs << res.source << std::endl;
+            validation_msgs << res.output;
+          }
+        } else {
+          validation_failed = true;
+          validation_msgs << "DXC executable not found. Cannot validate";
+        }
+        break;
+      }
+#endif  // TINT_BUILD_HLSL_WRITER
+#if TINT_BUILD_MSL_WRITER
+      case Format::kMsl: {
+#ifdef _WIN32
+        const char* default_xcrun_exe = "metal.exe";
+#else
+        const char* default_xcrun_exe = "xcrun";
+#endif
+        auto xcrun = tint::utils::Command::LookPath(options.xcrun_path.empty()
+                                                        ? default_xcrun_exe
+                                                        : options.xcrun_path);
+        if (xcrun.Found()) {
+          auto* w = static_cast<tint::writer::Text*>(writer.get());
+          auto msl = w->result();
+          auto res = tint::val::Msl(xcrun.Path(), msl);
+          if (res.failed) {
+            validation_failed = true;
+            validation_msgs << res.source << std::endl;
+            validation_msgs << res.output;
+          }
+        } else {
+          validation_failed = true;
+          validation_msgs << "xcrun executable not found. Cannot validate";
+        }
+        break;
+      }
+#endif  // TINT_BUILD_MSL_WRITER
+      default:
+        break;
     }
   }
 
+#if TINT_BUILD_SPV_WRITER
   if (options.format == Format::kSpvAsm) {
     auto* w = static_cast<tint::writer::spirv::Generator*>(writer.get());
     auto str = Disassemble(w->result());
@@ -841,12 +935,13 @@ int main(int argc, const char** argv) {
       return 1;
     }
   }
-  if (dawn_validation_failed) {
+#endif  // TINT_BUILD_SPV_WRITER
+
+  if (validation_failed) {
     std::cerr << std::endl << std::endl << "Validation Failure:" << std::endl;
-    std::cerr << stream.str();
+    std::cerr << validation_msgs.str();
     return 1;
   }
-#endif  // TINT_BUILD_SPV_WRITER
 
   if (options.format != Format::kSpvAsm && options.format != Format::kSpirv) {
     auto* w = static_cast<tint::writer::Text*>(writer.get());

src/BUILD.gn

@@ -208,6 +208,43 @@ tint_language_header("cldebuginfo100") {
   grammar_file = "${tint_spirv_headers_dir}/include/spirv/unified1/extinst.opencl.debuginfo.100.grammar.json"
 }
 
+###############################################################################
+# Helper library for IO operations
+# Only to be used by tests and sample executable
+###############################################################################
+source_set("tint_utils_io") {
+  sources = [
+    "utils/io/command.h",
+    "utils/io/tmpfile.h",
+  ]
+
+  if (is_linux || is_mac) {
+    sources += [ "utils/io/command_posix.cc" ]
+    sources += [ "utils/io/tmpfile_posix.cc" ]
+  } else if (is_win) {
+    sources += [ "utils/io/command_windows.cc" ]
+    sources += [ "utils/io/tmpfile_windows.cc" ]
+  } else {
+    sources += [ "utils/io/command_other.cc" ]
+    sources += [ "utils/io/tmpfile_other.cc" ]
+  }
+
+  public_deps = [ ":libtint_core_all_src" ]
+}
+
+###############################################################################
+# Helper library for validating generated shaders
+# As this depends on tint_utils_io, this is only to be used by tests and sample
+# executable
+###############################################################################
+source_set("tint_val") {
+  sources = [
+    "val/val.cc",
+    "val/val.h",
+  ]
+  public_deps = [ ":tint_utils_io" ]
+}
+
 ###############################################################################
 # Library - Tint core and optional modules of libtint
 ###############################################################################
@@ -464,9 +501,9 @@ libtint_source_set("libtint_core_all_src") {
     "sem/i32_type.cc",
     "sem/i32_type.h",
     "sem/info.h",
+    "sem/intrinsic.h",
     "sem/intrinsic_type.cc",
     "sem/intrinsic_type.h",
-    "sem/intrinsic.h",
     "sem/matrix_type.cc",
     "sem/matrix_type.h",
     "sem/multisampled_texture_type.cc",

src/CMakeLists.txt

@@ -449,6 +449,23 @@ if(${TINT_BUILD_HLSL_WRITER})
   )
 endif()
 
+## Tint IO utilities. Used by tint_val.
+add_library(tint_utils_io
+  utils/io/command_${TINT_OS_CC_SUFFIX}.cc
+  utils/io/command.h
+  utils/io/tmpfile_${TINT_OS_CC_SUFFIX}.cc
+  utils/io/tmpfile.h
+)
+tint_default_compile_options(tint_utils_io)
+
+## Tint validation utilities. Used by tests and the tint executable.
+add_library(tint_val
+  val/val.cc
+  val/val.h
+)
+tint_default_compile_options(tint_val)
+target_link_libraries(tint_val tint_utils_io)
+
 ## Tint library
 add_library(libtint ${TINT_LIB_SRCS})
 tint_default_compile_options(libtint)
@@ -603,16 +620,12 @@ if(${TINT_BUILD_TESTS})
     sem/type_manager_test.cc
     sem/u32_type_test.cc
     sem/vector_type_test.cc
-    utils/command_${TINT_OS_CC_SUFFIX}.cc
-    utils/command_test.cc
-    utils/command.h
     utils/get_or_create_test.cc
     utils/hash_test.cc
+    utils/io/command_test.cc
+    utils/io/tmpfile_test.cc
     utils/math_test.cc
     utils/scoped_assignment_test.cc
-    utils/tmpfile_${TINT_OS_CC_SUFFIX}.cc
-    utils/tmpfile_test.cc
-    utils/tmpfile.h
     utils/unique_vector_test.cc
     writer/append_vector_test.cc
     writer/float_to_string_test.cc
@@ -913,7 +926,7 @@ if(${TINT_BUILD_TESTS})
   ## Test executable
   target_include_directories(
       tint_unittests PRIVATE ${gmock_SOURCE_DIR}/include)
-  target_link_libraries(tint_unittests libtint gmock)
+  target_link_libraries(tint_unittests libtint gmock tint_val)
   tint_default_compile_options(tint_unittests)
 
   if(${TINT_BUILD_SPV_READER} OR ${TINT_BUILD_SPV_WRITER})

src/test_main.cc

@@ -14,7 +14,7 @@
 
 #include "gmock/gmock.h"
 #include "src/reader/spirv/parser_impl_test_helper.h"
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
 #include "src/writer/hlsl/test_helper.h"
 #include "src/writer/msl/test_helper.h"
 

src/utils/io/command.h

@@ -12,8 +12,8 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#ifndef SRC_UTILS_COMMAND_H_
+#ifndef SRC_UTILS_IO_COMMAND_H_
-#define SRC_UTILS_COMMAND_H_
+#define SRC_UTILS_IO_COMMAND_H_
 
 #include <string>
 #include <utility>
@@ -81,4 +81,4 @@ class Command {
 }  // namespace utils
 }  // namespace tint
 
-#endif  //  SRC_UTILS_COMMAND_H_
+#endif  //  SRC_UTILS_IO_COMMAND_H_

src/utils/io/command_other.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
 
 namespace tint {
 namespace utils {

src/utils/io/command_posix.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
 
 #include <sys/poll.h>
 #include <sys/stat.h>

src/utils/io/command_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
 
 #include "gtest/gtest.h"
 

src/utils/io/command_windows.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
 
 #define WIN32_LEAN_AND_MEAN 1
 #include <Windows.h>

src/utils/io/tmpfile.h

@@ -12,8 +12,8 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#ifndef SRC_UTILS_TMPFILE_H_
+#ifndef SRC_UTILS_IO_TMPFILE_H_
-#define SRC_UTILS_TMPFILE_H_
+#define SRC_UTILS_IO_TMPFILE_H_
 
 #include <sstream>
 #include <string>
@@ -73,4 +73,4 @@ class TmpFile {
 }  // namespace utils
 }  // namespace tint
 
-#endif  //  SRC_UTILS_TMPFILE_H_
+#endif  //  SRC_UTILS_IO_TMPFILE_H_

src/utils/io/tmpfile_other.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/tmpfile.h"
+#include "src/utils/io/tmpfile.h"
 
 namespace tint {
 namespace utils {

src/utils/io/tmpfile_posix.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/tmpfile.h"
+#include "src/utils/io/tmpfile.h"
 
 #include <unistd.h>
 #include <limits>

src/utils/io/tmpfile_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/tmpfile.h"
+#include "src/utils/io/tmpfile.h"
 
 #include <fstream>
 

src/utils/io/tmpfile_windows.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/utils/tmpfile.h"
+#include "src/utils/io/tmpfile.h"
 
 #include <stdio.h>
 #include <cstdio>

src/val/val.cc

@@ -0,0 +1,133 @@
+// Copyright 2021 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+#include "src/val/val.h"
+
+#include "src/ast/module.h"
+#include "src/program.h"
+#include "src/utils/io/command.h"
+#include "src/utils/io/tmpfile.h"
+
+namespace tint {
+namespace val {
+
+Result Hlsl(const std::string& dxc_path,
+            const std::string& source,
+            Program* program) {
+  Result result;
+
+  auto dxc = utils::Command(dxc_path);
+  if (!dxc.Found()) {
+    result.output = "DXC not found at '" + std::string(dxc_path) + "'";
+    result.failed = true;
+    return result;
+  }
+
+  result.source = source;
+
+  utils::TmpFile file;
+  file << source;
+
+  bool found_an_entrypoint = false;
+  for (auto* func : program->AST().Functions()) {
+    if (func->IsEntryPoint()) {
+      found_an_entrypoint = true;
+
+      const char* profile = "";
+
+      switch (func->pipeline_stage()) {
+        case ast::PipelineStage::kNone:
+          result.output = "Invalid PipelineStage";
+          result.failed = true;
+          return result;
+        case ast::PipelineStage::kVertex:
+          profile = "-T vs_6_0";
+          break;
+        case ast::PipelineStage::kFragment:
+          profile = "-T ps_6_0";
+          break;
+        case ast::PipelineStage::kCompute:
+          profile = "-T cs_6_0";
+          break;
+      }
+
+      auto name = program->Symbols().NameFor(func->symbol());
+      auto res = dxc(profile, "-E " + name, file.Path());
+      if (!res.out.empty()) {
+        if (!result.output.empty()) {
+          result.output += "\n";
+        }
+        result.output += res.out;
+      }
+      if (!res.err.empty()) {
+        if (!result.output.empty()) {
+          result.output += "\n";
+        }
+        result.output += res.err;
+      }
+      result.failed = (res.error_code != 0);
+    }
+  }
+
+  if (!found_an_entrypoint) {
+    result.output = "No entrypoint found";
+    result.failed = true;
+    return result;
+  }
+
+  return result;
+}
+
+Result Msl(const std::string& xcrun_path, const std::string& source) {
+  Result result;
+
+  auto xcrun = utils::Command(xcrun_path);
+  if (!xcrun.Found()) {
+    result.output = "xcrun not found at '" + std::string(xcrun_path) + "'";
+    result.failed = true;
+    return result;
+  }
+
+  result.source = source;
+
+  utils::TmpFile file(".metal");
+  file << result.source;
+
+#ifdef _WIN32
+  // On Windows, we should actually be running metal.exe from the Metal
+  // Developer Tools for Windows
+  auto res = xcrun("-x", "metal", "-c", "-o", "NUL", file.Path());
+#else
+  auto res =
+      xcrun("-sdk", "macosx", "metal", "-o", "/dev/null", "-c", file.Path());
+#endif
+  if (!res.out.empty()) {
+    if (!result.output.empty()) {
+      result.output += "\n";
+    }
+    result.output += res.out;
+  }
+  if (!res.err.empty()) {
+    if (!result.output.empty()) {
+      result.output += "\n";
+    }
+    result.output += res.err;
+  }
+  result.failed = (res.error_code != 0);
+
+  return result;
+}
+
+}  // namespace val
+}  // namespace tint

src/val/val.h

@@ -0,0 +1,58 @@
+// Copyright 2021 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+#ifndef SRC_VAL_VAL_H_
+#define SRC_VAL_VAL_H_
+
+#include <string>
+
+// Forward declarations
+namespace tint {
+class Program;
+}  // namespace tint
+
+namespace tint {
+namespace val {
+
+/// The return structure of Validate()
+struct Result {
+  /// True if validation passed
+  bool failed = false;
+  /// Output of DXC.
+  std::string output;
+  /// The generated source that was compiled
+  std::string source;
+};
+
+/// Hlsl attempts to compile the shader with DXC, verifying that the shader
+/// compiles successfully.
+/// @param dxc_path path to DXC
+/// @param source the generated HLSL source
+/// @param program the HLSL program
+/// @return the result of the compile
+Result Hlsl(const std::string& dxc_path,
+            const std::string& source,
+            Program* program);
+
+/// Msl attempts to compile the shader with the Metal Shader Compiler,
+/// verifying that the shader compiles successfully.
+/// @param xcrun_path path to xcrun
+/// @param source the generated MSL source
+/// @return the result of the compile
+Result Msl(const std::string& xcrun_path, const std::string& source);
+
+}  // namespace val
+}  // namespace tint
+
+#endif  // SRC_VAL_VAL_H_

src/writer/hlsl/test_helper.cc

@@ -14,9 +14,6 @@
 
 #include "src/writer/hlsl/test_helper.h"
 
-#include "src/utils/command.h"
-#include "src/utils/tmpfile.h"
-
 namespace tint {
 namespace writer {
 namespace hlsl {
@@ -31,81 +28,16 @@ void EnableHLSLValidation(const char* dxc) {
   dxc_path = dxc;
 }
 
-CompileResult Compile(Program* program, GeneratorImpl* generator) {
+val::Result Validate(Program* program, GeneratorImpl* generator) {
-  CompileResult result;
-
   if (!dxc_path) {
-    result.status = CompileResult::Status::kVerificationNotEnabled;
+    return val::Result{};
-    return result;
-  }
-
-  auto dxc = utils::Command(dxc_path);
-  if (!dxc.Found()) {
-    result.output = "DXC not found at '" + std::string(dxc_path) + "'";
-    result.status = CompileResult::Status::kFailed;
-    return result;
   }
 
   std::ostringstream hlsl;
   if (!generator->Generate(hlsl)) {
-    result.output = generator->error();
+    return {true, generator->error(), ""};
-    result.status = CompileResult::Status::kFailed;
-    return result;
   }
-  result.hlsl = hlsl.str();
+  return val::Hlsl(dxc_path, hlsl.str(), program);
-
-  utils::TmpFile file;
-  file << result.hlsl;
-
-  bool found_an_entrypoint = false;
-  for (auto* func : program->AST().Functions()) {
-    if (func->IsEntryPoint()) {
-      found_an_entrypoint = true;
-
-      const char* profile = "";
-
-      switch (func->pipeline_stage()) {
-        case ast::PipelineStage::kNone:
-          result.output = "Invalid PipelineStage";
-          result.status = CompileResult::Status::kFailed;
-          return result;
-        case ast::PipelineStage::kVertex:
-          profile = "-T vs_6_0";
-          break;
-        case ast::PipelineStage::kFragment:
-          profile = "-T ps_6_0";
-          break;
-        case ast::PipelineStage::kCompute:
-          profile = "-T cs_6_0";
-          break;
-      }
-
-      auto name = program->Symbols().NameFor(func->symbol());
-      auto res = dxc(profile, "-E " + name, file.Path());
-      if (!res.out.empty()) {
-        if (!result.output.empty()) {
-          result.output += "\n";
-        }
-        result.output += res.out;
-      }
-      if (!res.err.empty()) {
-        if (!result.output.empty()) {
-          result.output += "\n";
-        }
-        result.output += res.err;
-      }
-      result.status = (res.error_code == 0) ? CompileResult::Status::kSuccess
-                                            : CompileResult::Status::kFailed;
-    }
-  }
-
-  if (!found_an_entrypoint) {
-    result.output = "No entrypoint found";
-    result.status = CompileResult::Status::kFailed;
-    return result;
-  }
-
-  return result;
 }
 
 }  // namespace hlsl

src/writer/hlsl/test_helper.h

@@ -23,6 +23,7 @@
 #include "src/transform/hlsl.h"
 #include "src/transform/manager.h"
 #include "src/transform/renamer.h"
+#include "src/val/val.h"
 #include "src/writer/hlsl/generator_impl.h"
 
 namespace tint {
@@ -34,23 +35,11 @@ namespace hlsl {
 /// @param dxc_path the path to the DXC executable
 void EnableHLSLValidation(const char* dxc_path);
 
-/// The return structure of Compile()
+/// Validate attempts to compile the shader with DXC if found on PATH.
-struct CompileResult {
-  /// Status is an enumerator of status codes from Compile()
-  enum class Status { kSuccess, kFailed, kVerificationNotEnabled };
-  /// The resulting status of the compile
-  Status status;
-  /// Output of DXC.
-  std::string output;
-  /// The HLSL source that was compiled
-  std::string hlsl;
-};
-
-/// Compile attempts to compile the shader with DXC if found on PATH.
 /// @param program the HLSL program
 /// @param generator the HLSL generator
 /// @return the result of the compile
-CompileResult Compile(Program* program, GeneratorImpl* generator);
+val::Result Validate(Program* program, GeneratorImpl* generator);
 
 /// Helper class for testing
 template <typename BODY>
@@ -122,10 +111,10 @@ class TestHelperBase : public BODY, public ProgramBuilder {
   /// If DXC finds problems the test will fail.
   /// If DXC is not on `PATH` then Validate() does nothing.
   void Validate() const {
-    auto res = Compile(program.get(), gen_.get());
+    auto res = hlsl::Validate(program.get(), gen_.get());
-    if (res.status == CompileResult::Status::kFailed) {
+    if (res.failed) {
       FAIL() << "HLSL Validation failed.\n\n"
-             << res.hlsl << "\n\n"
+             << res.source << "\n\n"
              << res.output;
     }
   }

src/writer/msl/test_helper.cc

@@ -14,8 +14,8 @@
 
 #include "src/writer/msl/test_helper.h"
 
-#include "src/utils/command.h"
+#include "src/utils/io/command.h"
-#include "src/utils/tmpfile.h"
+#include "src/utils/io/tmpfile.h"
 
 namespace tint {
 namespace writer {
@@ -31,50 +31,16 @@ void EnableMSLValidation(const char* xcrun) {
   xcrun_path = xcrun;
 }
 
-CompileResult Compile(Program* program) {
+val::Result Validate(Program* program) {
-  CompileResult result;
-
   if (!xcrun_path) {
-    result.status = CompileResult::Status::kVerificationNotEnabled;
+    return val::Result{};
-    return result;
-  }
-
-  auto xcrun = utils::Command(xcrun_path);
-  if (!xcrun.Found()) {
-    result.output = "xcrun not found at '" + std::string(xcrun_path) + "'";
-    result.status = CompileResult::Status::kFailed;
-    return result;
   }
 
   auto gen = std::make_unique<GeneratorImpl>(program);
   if (!gen->Generate()) {
-    result.output = gen->error();
+    return {true, gen->error(), ""};
-    result.status = CompileResult::Status::kFailed;
-    return result;
   }
-  result.msl = gen->result();
+  return val::Msl(xcrun_path, gen->result());
-
-  utils::TmpFile file(".metal");
-  file << result.msl;
-
-  auto xcrun_res =
-      xcrun("-sdk", "macosx", "metal", "-o", "/dev/null", "-c", file.Path());
-  if (!xcrun_res.out.empty()) {
-    if (!result.output.empty()) {
-      result.output += "\n";
-    }
-    result.output += xcrun_res.out;
-  }
-  if (!xcrun_res.err.empty()) {
-    if (!result.output.empty()) {
-      result.output += "\n";
-    }
-    result.output += xcrun_res.err;
-  }
-  result.status = (xcrun_res.error_code == 0) ? CompileResult::Status::kSuccess
-                                              : CompileResult::Status::kFailed;
-
-  return result;
 }
 
 }  // namespace msl

src/writer/msl/test_helper.h

@@ -22,6 +22,7 @@
 #include "gtest/gtest.h"
 #include "src/program_builder.h"
 #include "src/transform/msl.h"
+#include "src/val/val.h"
 #include "src/writer/msl/generator_impl.h"
 
 namespace tint {
@@ -33,22 +34,10 @@ namespace msl {
 /// @param xcrun_path the path to the `xcrun` executable
 void EnableMSLValidation(const char* xcrun_path);
 
-/// The return structure of Compile()
+/// Validate attempts to compile the shader with DXC if found on PATH.
-struct CompileResult {
-  /// Status is an enumerator of status codes from Compile()
-  enum class Status { kSuccess, kFailed, kVerificationNotEnabled };
-  /// The resulting status of the compilation
-  Status status;
-  /// Output of the Metal compiler
-  std::string output;
-  /// The MSL source that was compiled
-  std::string msl;
-};
-
-/// Compile attempts to compile the shader with xcrun if found on PATH.
 /// @param program the MSL program
 /// @return the result of the compile
-CompileResult Compile(Program* program);
+val::Result Validate(Program* program);
 
 /// Helper class for testing
 template <typename BASE>
@@ -115,9 +104,11 @@ class TestHelperBase : public BASE, public ProgramBuilder {
   /// This function does nothing, if the Metal compiler path has not been
   /// configured by calling `EnableMSLValidation()`.
   void Validate() {
-    auto res = Compile(program.get());
+    auto res = msl::Validate(program.get());
-    if (res.status == CompileResult::Status::kFailed) {
+    if (res.failed) {
-      FAIL() << "MSL Validation failed.\n\n" << res.msl << "\n\n" << res.output;
+      FAIL() << "MSL Validation failed.\n\n"
+             << res.source << "\n\n"
+             << res.output;
     }
   }
 

test/BUILD.gn

@@ -90,11 +90,11 @@ source_set("tint_unittests_main") {
     sources = [ "../src/test_main.cc" ]
     configs += [ ":tint_unittests_config" ]
     deps += [
-      ":tint_test_helpers",
       ":tint_unittests_hlsl_writer_src",
       ":tint_unittests_msl_writer_src",
       ":tint_unittests_spv_reader_src",
       "${tint_root_dir}/src:libtint",
+      "${tint_root_dir}/src:tint_val",
     ]
   }
 }
@@ -114,36 +114,6 @@ config("tint_unittests_config") {
   ]
 }
 
-source_set("tint_test_helpers") {
-  testonly = true
-
-  sources = [
-    "../src/ast/intrinsic_texture_helper_test.cc",
-    "../src/ast/intrinsic_texture_helper_test.h",
-    "../src/transform/test_helper.h",
-    "../src/utils/command.h",
-    "../src/utils/tmpfile.h",
-  ]
-
-  if (is_linux || is_mac) {
-    sources += [ "../src/utils/command_posix.cc" ]
-    sources += [ "../src/utils/tmpfile_posix.cc" ]
-  } else if (is_win) {
-    sources += [ "../src/utils/command_windows.cc" ]
-    sources += [ "../src/utils/tmpfile_windows.cc" ]
-  } else {
-    sources += [ "../src/utils/command_other.cc" ]
-    sources += [ "../src/utils/tmpfile_other.cc" ]
-  }
-
-  configs += [ ":tint_unittests_config" ]
-
-  public_deps = [
-    ":gmock_and_gtest",
-    "${tint_root_dir}/src:libtint",
-  ]
-}
-
 template("tint_unittests_source_set") {
   source_set(target_name) {
     forward_variables_from(invoker, "*", [ "configs" ])
@@ -162,7 +132,11 @@ template("tint_unittests_source_set") {
     if (!defined(invoker.deps)) {
       deps = []
     }
-    deps += [ ":tint_test_helpers" ]
+    deps += [
+      ":gmock_and_gtest",
+      "${tint_root_dir}/src:libtint",
+      "${tint_root_dir}/src:tint_val",
+    ]
   }
 }
 
@@ -203,6 +177,8 @@ tint_unittests_source_set("tint_unittests_core_src") {
     "../src/ast/identifier_expression_test.cc",
     "../src/ast/if_statement_test.cc",
     "../src/ast/int_literal_test.cc",
+    "../src/ast/intrinsic_texture_helper_test.cc",
+    "../src/ast/intrinsic_texture_helper_test.h",
     "../src/ast/location_decoration_test.cc",
     "../src/ast/loop_statement_test.cc",
     "../src/ast/matrix_test.cc",
@@ -307,15 +283,16 @@ tint_unittests_source_set("tint_unittests_core_src") {
     "../src/transform/renamer_test.cc",
     "../src/transform/simplify_test.cc",
     "../src/transform/single_entry_point_test.cc",
+    "../src/transform/test_helper.h",
     "../src/transform/transform_test.cc",
     "../src/transform/var_for_dynamic_index_test.cc",
     "../src/transform/vertex_pulling_test.cc",
-    "../src/utils/command_test.cc",
     "../src/utils/get_or_create_test.cc",
     "../src/utils/hash_test.cc",
+    "../src/utils/io/command_test.cc",
+    "../src/utils/io/tmpfile_test.cc",
     "../src/utils/math_test.cc",
     "../src/utils/scoped_assignment_test.cc",
-    "../src/utils/tmpfile_test.cc",
     "../src/utils/unique_vector_test.cc",
     "../src/writer/append_vector_test.cc",
     "../src/writer/float_to_string_test.cc",
@@ -365,7 +342,10 @@ tint_unittests_source_set("tint_unittests_spv_reader_src") {
     "../src/reader/spirv/usage_test.cc",
   ]
 
-  deps = [ "${tint_root_dir}/src:libtint_spv_reader_src" ]
+  deps = [
+    ":tint_unittests_core_src",
+    "${tint_root_dir}/src:libtint_spv_reader_src",
+  ]
 }
 
 tint_unittests_source_set("tint_unittests_spv_writer_src") {
@@ -406,6 +386,7 @@ tint_unittests_source_set("tint_unittests_spv_writer_src") {
   ]
 
   deps = [
+    ":tint_unittests_core_src",
     "${tint_root_dir}/src:libtint_spv_writer_src",
     "${tint_spirv_tools_dir}/:spvtools",
   ]
@@ -489,7 +470,10 @@ tint_unittests_source_set("tint_unittests_wgsl_reader_src") {
     "../src/reader/wgsl/token_test.cc",
   ]
 
-  deps = [ "${tint_root_dir}/src:libtint_wgsl_reader_src" ]
+  deps = [
+    ":tint_unittests_core_src",
+    "${tint_root_dir}/src:libtint_wgsl_reader_src",
+  ]
 }
 
 tint_unittests_source_set("tint_unittests_wgsl_writer_src") {
@@ -525,7 +509,10 @@ tint_unittests_source_set("tint_unittests_wgsl_writer_src") {
     "../src/writer/wgsl/test_helper.h",
   ]
 
-  deps = [ "${tint_root_dir}/src:libtint_wgsl_writer_src" ]
+  deps = [
+    ":tint_unittests_core_src",
+    "${tint_root_dir}/src:libtint_wgsl_writer_src",
+  ]
 }
 
 tint_unittests_source_set("tint_unittests_msl_writer_src") {
@@ -563,7 +550,10 @@ tint_unittests_source_set("tint_unittests_msl_writer_src") {
     "../src/writer/msl/test_helper.h",
   ]
 
-  deps = [ "${tint_root_dir}/src:libtint_msl_writer_src" ]
+  deps = [
+    ":tint_unittests_core_src",
+    "${tint_root_dir}/src:libtint_msl_writer_src",
+  ]
 }
 
 tint_unittests_source_set("tint_unittests_hlsl_writer_src") {
@@ -603,7 +593,10 @@ tint_unittests_source_set("tint_unittests_hlsl_writer_src") {
     "../src/writer/hlsl/test_helper.h",
   ]
 
-  deps = [ "${tint_root_dir}/src:libtint_hlsl_writer_src" ]
+  deps = [
+    ":tint_unittests_core_src",
+    "${tint_root_dir}/src:libtint_hlsl_writer_src",
+  ]
 }
 
 source_set("tint_unittests_src") {

test/bug/tint/221.wgsl

@@ -6,7 +6,7 @@ struct Buf{
   data : Arr;
 };
 
-[[group(0), binding (0)]] var<storage>  b : [[access(read)]] Buf;
+[[group(0), binding (0)]] var<storage>  b : [[access(read_write)]] Buf;
 
 [[stage(compute)]]
 fn main() {

test/bug/tint/221.wgsl.expected.hlsl

@@ -1,5 +1,5 @@
 
-ByteAddressBuffer b : register(t0, space0);
+RWByteAddressBuffer b : register(u0, space0);
 
 [numthreads(1, 1, 1)]
 void main() {

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

@@ -8,7 +8,7 @@ struct Buf {
   /* 0x0004 */ uint data[50];
 };
 
-kernel void tint_symbol(const device Buf& b [[buffer(0)]]) {
+kernel void tint_symbol(device Buf& b [[buffer(0)]]) {
   uint i = 0u;
   {
     bool tint_msl_is_first_1 = true;

test/bug/tint/221.wgsl.expected.spvasm

@@ -17,7 +17,6 @@
                OpMemberDecorate %Buf 0 Offset 0
                OpMemberDecorate %Buf 1 Offset 4
                OpDecorate %_arr_uint_uint_50 ArrayStride 4
-               OpDecorate %b NonWritable
                OpDecorate %b DescriptorSet 0
                OpDecorate %b Binding 0
        %uint = OpTypeInt 32 0

test/bug/tint/221.wgsl.expected.wgsl

@@ -6,7 +6,7 @@ struct Buf {
   data : Arr;
 };
 
-[[group(0), binding(0)]] var<storage> b : [[access(read)]] Buf;
+[[group(0), binding(0)]] var<storage> b : [[access(read_write)]] Buf;
 
 [[stage(compute)]]
 fn main() {

test/test-all.sh

@@ -20,7 +20,7 @@ SCRIPT_DIR="$( cd "$( dirname "${BASH_SOURCE[0]}")" >/dev/null 2>&1 && pwd )"
 
 function usage() {
     echo "test-all.sh is a simple wrapper around <tint>/tools/test-runner that"
-    echo "injects the <tint>/tools directory as the second command line argument"
+    echo "injects the <tint>/test directory as the second command line argument"
     echo
     echo "Usage of <tint>/tools/test-runner:"
     "${SCRIPT_DIR}/../tools/test-runner" --help
@@ -41,4 +41,4 @@ if [ ! -x "$TINT" ]; then
     exit 1
 fi
 
-"${SCRIPT_DIR}/../tools/test-runner" ${@:2} "${TINT}" "${SCRIPT_DIR}"
+"${SCRIPT_DIR}/../tools/test-runner" "${@:2}" "${TINT}" "${SCRIPT_DIR}"

tools/src/cmd/test-runner/main.go

@@ -68,11 +68,13 @@ optional flags:`)
 }
 
 func run() error {
-	var formatList, filter string
+	var formatList, filter, dxcPath, xcrunPath string
 	numCPU := runtime.NumCPU()
 	generateExpected := false
 	flag.StringVar(&formatList, "format", "all", "comma separated list of formats to emit. Possible values are: all, wgsl, spvasm, msl, hlsl")
 	flag.StringVar(&filter, "filter", "**.wgsl, **.spvasm, **.spv", "comma separated list of glob patterns for test files")
+	flag.StringVar(&dxcPath, "dxc", "", "path to DXC executable for validating HLSL output")
+	flag.StringVar(&xcrunPath, "xcrun", "", "path to xcrun executable for validating MSL output")
 	flag.BoolVar(&generateExpected, "generate-expected", false, "create or update all expected outputs")
 	flag.IntVar(&numCPU, "j", numCPU, "maximum number of concurrent threads to run tests")
 	flag.Usage = showUsage
@@ -147,6 +149,46 @@ func run() error {
 		}
 	}
 
+	default_msl_exe := "xcrun"
+	if runtime.GOOS == "windows" {
+		default_msl_exe = "metal.exe"
+	}
+
+	// If explicit verification compilers have been specified, check they exist.
+	// Otherwise, look on PATH for them, but don't error if they cannot be found.
+	for _, tool := range []struct {
+		name string
+		lang string
+		path *string
+	}{
+		{"dxc", "hlsl", &dxcPath},
+		{default_msl_exe, "msl", &xcrunPath},
+	} {
+		if *tool.path == "" {
+			p, err := exec.LookPath(tool.name)
+			if err == nil && fileutils.IsExe(p) {
+				*tool.path = p
+			}
+		} else if !fileutils.IsExe(*tool.path) {
+			return fmt.Errorf("%v not found at '%v'", tool.name, *tool.path)
+		}
+
+		color.Set(color.FgCyan)
+		fmt.Printf("%-4s", tool.lang)
+		color.Unset()
+		fmt.Printf(" validation ")
+		if *tool.path == "" {
+			color.Set(color.FgRed)
+			fmt.Printf("DISABLED")
+		} else {
+			color.Set(color.FgGreen)
+			fmt.Printf("ENABLED")
+		}
+		color.Unset()
+		fmt.Println()
+	}
+	fmt.Println()
+
 	results := make([]map[outputFormat]*status, len(files))
 	jobs := make(chan job, 256)
 
@@ -157,7 +199,7 @@ func run() error {
 		go func() {
 			defer wg.Done()
 			for job := range jobs {
-				job.run(exe, generateExpected)
+				job.run(exe, dxcPath, xcrunPath, generateExpected)
 			}
 		}()
 	}
@@ -313,7 +355,7 @@ type job struct {
 	result *status
 }
 
-func (j job) run(exe string, generateExpected bool) {
+func (j job) run(exe, dxcPath, xcrunPath string, generateExpected bool) {
 	// Is there an expected output?
 	expected := loadExpectedFile(j.file, j.format)
 	if strings.HasPrefix(expected, "SKIP") { // Special SKIP token
@@ -321,9 +363,28 @@ func (j job) run(exe string, generateExpected bool) {
 		return
 	}
 
+	expected = strings.ReplaceAll(expected, "\r\n", "\n")
+
+	args := []string{j.file, "--format", string(j.format)}
+
+	// Can we validate?
+	switch j.format {
+	case spvasm:
+		args = append(args, "--validate") // spirv-val is statically linked, always available
+	case hlsl:
+		if dxcPath != "" {
+			args = append(args, "--dxc", dxcPath)
+		}
+	case msl:
+		if xcrunPath != "" {
+			args = append(args, "--xcrun", xcrunPath)
+		}
+	}
+
 	// Invoke the compiler...
 	var err error
-	if ok, out := invoke(exe, j.file, "--format", string(j.format), "--dawn-validation"); ok {
+	if ok, out := invoke(exe, args...); ok {
+		out = strings.ReplaceAll(out, "\r\n", "\n")
 		if generateExpected {
 			// If --generate-expected was passed, write out the output
 			err = saveExpectedFile(j.file, j.format, out)

tools/src/fileutils/fileutils_windows.go

@@ -0,0 +1,21 @@
+// Copyright 2021 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+// Package fileutils contains utility functions for files
+package fileutils
+
+// IsExe returns true if the file at path is an executable
+func IsExe(path string) bool {
+	return true
+}