Import Tint changes from Dawn
Changes:
- 46bd2e101797c5b016303a54b8cee001567b3fc8 Move remote-compile to src/tint/cmd by dan sinclair <dsinclair@chromium.org>
- 67f09927c0c06bc0f52e816d2420533e4433b1f2 Add TextureBuiltinsFromUniform transform by Shrek Shao <shrekshao@google.com>
GitOrigin-RevId: 46bd2e101797c5b016303a54b8cee001567b3fc8
Change-Id: I222b75bd78ee3b7f3189c00b275279fdb9feff6e
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/142200
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
diff --git a/include/tint/tint.h b/include/tint/tint.h
index c8f6581..899de9a 100644
--- a/include/tint/tint.h
+++ b/include/tint/tint.h
@@ -25,6 +25,7 @@
#include "src/tint/api/options/array_length_from_uniform.h"
#include "src/tint/api/options/binding_remapper.h"
#include "src/tint/api/options/external_texture.h"
+#include "src/tint/api/options/texture_builtins_from_uniform.h"
#include "src/tint/api/tint.h"
#include "src/tint/lang/core/type/manager.h"
#include "src/tint/lang/wgsl/ast/transform/first_index_offset.h"
diff --git a/src/tint/BUILD.gn b/src/tint/BUILD.gn
index b97374d..ea62726 100644
--- a/src/tint/BUILD.gn
+++ b/src/tint/BUILD.gn
@@ -199,6 +199,10 @@
}
}
+source_set("winsock") {
+ libs = [ "ws2_32.lib" ]
+}
+
###############################################################################
# Aliases.
###############################################################################
diff --git a/src/tint/CMakeLists.txt b/src/tint/CMakeLists.txt
index 1720ead..ef4ce00 100644
--- a/src/tint/CMakeLists.txt
+++ b/src/tint/CMakeLists.txt
@@ -448,6 +448,8 @@
target_link_libraries(${TARGET} PRIVATE
glslang-default-resource-limits
)
+ elseif(${DEPENDENCY} STREQUAL "winsock")
+ target_link_libraries(${TARGET} PRIVATE ws2_32)
else()
message(FATAL_ERROR "unhandled external dependency ${DEPENDENCY}")
endif()
@@ -492,6 +494,9 @@
################################################################################
include("BUILD.cmake")
+find_package(Threads REQUIRED)
+target_link_libraries(tint_cmd_remote_compile_cmd PRIVATE Threads::Threads)
+
# If we're building on mac / ios and we have CoreGraphics, then we can use the
# metal API to validate our shaders. This is roughly 4x faster than invoking
# the metal shader compiler executable.
@@ -500,6 +505,7 @@
if(LIB_CORE_GRAPHICS)
target_sources("tint_lang_msl_validate" PRIVATE "lang/msl/validate/msl_metal.mm")
+ target_compile_definitions("tint_cmd_remote_compile_cmd" PRIVATE "-DTINT_ENABLE_MSL_VALIDATION_USING_METAL_API=1")
target_compile_definitions("tint_lang_msl_validate" PUBLIC "-DTINT_ENABLE_MSL_VALIDATION_USING_METAL_API=1")
target_compile_options("tint_lang_msl_validate" PRIVATE "-fmodules" "-fcxx-modules")
target_link_options("tint_lang_msl_validate" PUBLIC "-framework" "CoreGraphics")
diff --git a/src/tint/api/options/BUILD.cmake b/src/tint/api/options/BUILD.cmake
index 2edad38..8117549 100644
--- a/src/tint/api/options/BUILD.cmake
+++ b/src/tint/api/options/BUILD.cmake
@@ -29,6 +29,7 @@
api/options/binding_remapper.h
api/options/external_texture.h
api/options/options.cc
+ api/options/texture_builtins_from_uniform.h
)
tint_target_add_dependencies(tint_api_options lib
diff --git a/src/tint/api/options/BUILD.gn b/src/tint/api/options/BUILD.gn
index e24f6f1..9608ec4 100644
--- a/src/tint/api/options/BUILD.gn
+++ b/src/tint/api/options/BUILD.gn
@@ -30,6 +30,7 @@
"binding_remapper.h",
"external_texture.h",
"options.cc",
+ "texture_builtins_from_uniform.h",
]
deps = [
"${tint_src_dir}/api/common",
diff --git a/src/tint/api/options/texture_builtins_from_uniform.h b/src/tint/api/options/texture_builtins_from_uniform.h
new file mode 100644
index 0000000..4dc2eff
--- /dev/null
+++ b/src/tint/api/options/texture_builtins_from_uniform.h
@@ -0,0 +1,52 @@
+// Copyright 2023 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_TINT_API_OPTIONS_TEXTURE_BUILTINS_FROM_UNIFORM_H_
+#define SRC_TINT_API_OPTIONS_TEXTURE_BUILTINS_FROM_UNIFORM_H_
+
+#include <unordered_map>
+#include <utility>
+
+#include "src/tint/api/common/binding_point.h"
+#include "src/tint/utils/reflection/reflection.h"
+
+namespace tint {
+
+/// Options used to specify a mapping of binding points to indices into a UBO
+/// from which to load buffer sizes.
+struct TextureBuiltinsFromUniformOptions {
+ /// Indicate the type of field for each entry to push.
+ enum class Field {
+ /// The number of mip levels of the bonnd texture view.
+ TextureNumLevels,
+ /// The number of samples per texel of the bound multipsampled texture.
+ TextureNumSamples,
+ };
+
+ /// Records the field and the byte offset of the data to push in the internal uniform buffer.
+ using FieldAndOffset = std::pair<Field, uint32_t>;
+ /// Maps from binding point to data entry with the information to populate the data.
+ using BindingPointToFieldAndOffset = std::unordered_map<BindingPoint, FieldAndOffset>;
+
+ /// The binding point to use to generate a uniform buffer from which to read
+ /// buffer sizes.
+ BindingPoint ubo_binding = {};
+
+ /// Reflect the fields of this class so that it can be used by tint::ForeachField()
+ TINT_REFLECT(ubo_binding);
+};
+
+} // namespace tint
+
+#endif // SRC_TINT_API_OPTIONS_TEXTURE_BUILTINS_FROM_UNIFORM_H_
diff --git a/src/tint/cmd/BUILD.cmake b/src/tint/cmd/BUILD.cmake
index f547319..cbad9b9 100644
--- a/src/tint/cmd/BUILD.cmake
+++ b/src/tint/cmd/BUILD.cmake
@@ -24,5 +24,6 @@
include(cmd/common/BUILD.cmake)
include(cmd/info/BUILD.cmake)
include(cmd/loopy/BUILD.cmake)
+include(cmd/remote_compile/BUILD.cmake)
include(cmd/test/BUILD.cmake)
include(cmd/tint/BUILD.cmake)
diff --git a/src/tint/cmd/remote_compile/BUILD.cfg b/src/tint/cmd/remote_compile/BUILD.cfg
new file mode 100644
index 0000000..046b778
--- /dev/null
+++ b/src/tint/cmd/remote_compile/BUILD.cfg
@@ -0,0 +1,3 @@
+{
+ "cmd": { "OutputName": "tint_remote_compile" }
+}
diff --git a/src/tint/cmd/remote_compile/BUILD.cmake b/src/tint/cmd/remote_compile/BUILD.cmake
new file mode 100644
index 0000000..a92b5bd
--- /dev/null
+++ b/src/tint/cmd/remote_compile/BUILD.cmake
@@ -0,0 +1,45 @@
+# Copyright 2023 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.
+
+################################################################################
+# File generated by tools/src/cmd/gen
+# using the template:
+# tools/src/cmd/gen/build/BUILD.cmake.tmpl
+#
+# Do not modify this file directly
+################################################################################
+
+################################################################################
+# Target: tint_cmd_remote_compile_cmd
+# Kind: cmd
+################################################################################
+tint_add_target(tint_cmd_remote_compile_cmd cmd
+ cmd/remote_compile/main.cc
+)
+
+tint_target_add_dependencies(tint_cmd_remote_compile_cmd cmd
+ tint_lang_wgsl_ast
+ tint_utils_macros
+ tint_utils_socket
+ tint_utils_text
+ tint_utils_traits
+)
+
+if(TINT_BUILD_MSL_WRITER)
+ tint_target_add_dependencies(tint_cmd_remote_compile_cmd cmd
+ tint_lang_msl_validate
+ )
+endif(TINT_BUILD_MSL_WRITER)
+
+tint_target_set_output_name(tint_cmd_remote_compile_cmd cmd "tint_remote_compile")
diff --git a/src/tint/cmd/remote_compile/BUILD.gn b/src/tint/cmd/remote_compile/BUILD.gn
new file mode 100644
index 0000000..033196a
--- /dev/null
+++ b/src/tint/cmd/remote_compile/BUILD.gn
@@ -0,0 +1,41 @@
+# Copyright 2023 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.
+
+################################################################################
+# File generated by tools/src/cmd/gen
+# using the template:
+# tools/src/cmd/gen/build/BUILD.gn.tmpl
+#
+# Do not modify this file directly
+################################################################################
+
+import("../../../../scripts/tint_overrides_with_defaults.gni")
+
+import("${tint_src_dir}/tint.gni")
+
+executable("remote_compile") {
+ output_name = "tint_remote_compile"
+ sources = [ "main.cc" ]
+ deps = [
+ "${tint_src_dir}/lang/wgsl/ast",
+ "${tint_src_dir}/utils/macros",
+ "${tint_src_dir}/utils/socket",
+ "${tint_src_dir}/utils/text",
+ "${tint_src_dir}/utils/traits",
+ ]
+
+ if (tint_build_msl_writer) {
+ deps += [ "${tint_src_dir}/lang/msl/validate" ]
+ }
+}
diff --git a/src/tint/cmd/remote_compile/main.cc b/src/tint/cmd/remote_compile/main.cc
new file mode 100644
index 0000000..a13870e
--- /dev/null
+++ b/src/tint/cmd/remote_compile/main.cc
@@ -0,0 +1,494 @@
+// 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
+//
+// https://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 <stdio.h>
+#include <stdlib.h>
+#include <fstream>
+#include <iostream>
+#include <regex>
+#include <sstream>
+#include <string>
+#include <thread>
+#include <type_traits>
+#include <vector>
+
+#include "src/tint/lang/msl/validate/val.h"
+#include "src/tint/utils/macros/compiler.h"
+#include "src/tint/utils/socket/socket.h"
+
+namespace {
+
+#if 0
+#define DEBUG(msg, ...) printf(msg "\n", ##__VA_ARGS__)
+#else
+#define DEBUG(...)
+#endif
+
+/// The return structure of a compile function
+struct CompileResult {
+ /// True if shader compiled
+ bool success = false;
+ /// Output of the compiler
+ std::string output;
+};
+
+/// Print the tool usage, and exit with 1.
+[[noreturn]] void ShowUsage() {
+ const char* name = "tint-remote-compile";
+ printf(R"(%s is a tool for compiling a shader on a remote machine
+
+usage as server:
+ %s -s [-p port-number]
+
+usage as client:
+ %s [-p port-number] [server-address] shader-file-path
+
+ [server-address] can be omitted if the TINT_REMOTE_COMPILE_ADDRESS environment
+ variable is set.
+ Alternatively, you can pass xcrun arguments so %s can be used as a
+ drop-in replacement.
+)",
+ name, name, name, name);
+ exit(1);
+}
+
+/// The protocol version code. Bump each time the protocol changes
+constexpr uint32_t kProtocolVersion = 1;
+
+/// Supported shader source languages
+enum SourceLanguage {
+ MSL,
+};
+
+/// Stream is a serialization wrapper around a socket
+struct Stream {
+ /// The underlying socket
+ Socket* const socket;
+ /// Error state
+ std::string error;
+
+ /// Writes a uint32_t to the socket
+ Stream operator<<(uint32_t v) {
+ if (error.empty()) {
+ Write(&v, sizeof(v));
+ }
+ return *this;
+ }
+
+ /// Reads a uint32_t from the socket
+ Stream operator>>(uint32_t& v) {
+ if (error.empty()) {
+ Read(&v, sizeof(v));
+ }
+ return *this;
+ }
+
+ /// Writes a std::string to the socket
+ Stream operator<<(const std::string& v) {
+ if (error.empty()) {
+ uint32_t count = static_cast<uint32_t>(v.size());
+ *this << count;
+ if (count) {
+ Write(v.data(), count);
+ }
+ }
+ return *this;
+ }
+
+ /// Reads a std::string from the socket
+ Stream operator>>(std::string& v) {
+ uint32_t count = 0;
+ *this >> count;
+ if (count) {
+ std::vector<char> buf(count);
+ if (Read(buf.data(), count)) {
+ v = std::string(buf.data(), buf.size());
+ }
+ } else {
+ v.clear();
+ }
+ return *this;
+ }
+
+ /// Writes an enum value to the socket
+ template <typename T>
+ std::enable_if_t<std::is_enum<T>::value, Stream> operator<<(T e) {
+ return *this << static_cast<uint32_t>(e);
+ }
+
+ /// Reads an enum value from the socket
+ template <typename T>
+ std::enable_if_t<std::is_enum<T>::value, Stream> operator>>(T& e) {
+ uint32_t v;
+ *this >> v;
+ e = static_cast<T>(v);
+ return *this;
+ }
+
+ private:
+ bool Write(const void* data, size_t size) {
+ if (error.empty()) {
+ if (!socket->Write(data, size)) {
+ error = "Socket::Write() failed";
+ }
+ }
+ return error.empty();
+ }
+
+ bool Read(void* data, size_t size) {
+ auto buf = reinterpret_cast<uint8_t*>(data);
+ while (size > 0 && error.empty()) {
+ if (auto n = socket->Read(buf, size)) {
+ if (n > size) {
+ error = "Socket::Read() returned more bytes than requested";
+ return false;
+ }
+ size -= n;
+ buf += n;
+ } else {
+ error = "Socket::Read() failed";
+ }
+ }
+ return error.empty();
+ }
+};
+
+////////////////////////////////////////////////////////////////////////////////
+// Messages
+////////////////////////////////////////////////////////////////////////////////
+
+/// Base class for all messages
+struct Message {
+ /// The type of the message
+ enum class Type {
+ ConnectionRequest,
+ ConnectionResponse,
+ CompileRequest,
+ CompileResponse,
+ };
+
+ explicit Message(Type ty) : type(ty) {}
+
+ const Type type;
+};
+
+struct ConnectionResponse : Message { // Server -> Client
+ ConnectionResponse() : Message(Type::ConnectionResponse) {}
+
+ template <typename T>
+ void Serialize(T&& f) {
+ f(error);
+ }
+
+ std::string error;
+};
+
+struct ConnectionRequest : Message { // Client -> Server
+ using Response = ConnectionResponse;
+
+ explicit ConnectionRequest(uint32_t proto_ver = kProtocolVersion)
+ : Message(Type::ConnectionRequest), protocol_version(proto_ver) {}
+
+ template <typename T>
+ void Serialize(T&& f) {
+ f(protocol_version);
+ }
+
+ uint32_t protocol_version;
+};
+
+struct CompileResponse : Message { // Server -> Client
+ CompileResponse() : Message(Type::CompileResponse) {}
+
+ template <typename T>
+ void Serialize(T&& f) {
+ f(error);
+ }
+
+ std::string error;
+};
+
+struct CompileRequest : Message { // Client -> Server
+ using Response = CompileResponse;
+
+ CompileRequest() : Message(Type::CompileRequest) {}
+ CompileRequest(SourceLanguage lang, int ver_major, int ver_minor, std::string src)
+ : Message(Type::CompileRequest),
+ language(lang),
+ version_major(uint32_t(ver_major)),
+ version_minor(uint32_t(ver_minor)),
+ source(src) {}
+
+ template <typename T>
+ void Serialize(T&& f) {
+ f(language);
+ f(source);
+ f(version_major);
+ f(version_minor);
+ }
+
+ SourceLanguage language = SourceLanguage::MSL;
+ uint32_t version_major = 0;
+ uint32_t version_minor = 0;
+ std::string source;
+};
+
+/// Writes the message `m` to the stream `s`
+template <typename MESSAGE>
+std::enable_if_t<std::is_base_of<Message, MESSAGE>::value, Stream>& operator<<(Stream& s,
+ const MESSAGE& m) {
+ s << m.type;
+ const_cast<MESSAGE&>(m).Serialize([&s](const auto& value) { s << value; });
+ return s;
+}
+
+/// Reads the message `m` from the stream `s`
+template <typename MESSAGE>
+std::enable_if_t<std::is_base_of<Message, MESSAGE>::value, Stream>& operator>>(Stream& s,
+ MESSAGE& m) {
+ Message::Type ty;
+ s >> ty;
+ if (s.error.empty()) {
+ if (ty == m.type) {
+ m.Serialize([&s](auto& value) { s >> value; });
+ } else {
+ std::stringstream ss;
+ ss << "expected message type " << static_cast<int>(m.type) << ", got "
+ << static_cast<int>(ty);
+ s.error = ss.str();
+ }
+ }
+ return s;
+}
+
+/// Writes the request message `req` to the stream `s`, then reads and returns
+/// the response message from the same stream.
+template <typename REQUEST, typename RESPONSE = typename REQUEST::Response>
+RESPONSE Send(Stream& s, const REQUEST& req) {
+ s << req;
+ if (s.error.empty()) {
+ RESPONSE resp;
+ s >> resp;
+ if (s.error.empty()) {
+ return resp;
+ }
+ }
+ return {};
+}
+
+} // namespace
+
+bool RunServer(std::string port);
+bool RunClient(std::string address,
+ std::string port,
+ std::string file,
+ int version_major,
+ int version_minor);
+
+int main(int argc, char* argv[]) {
+ bool run_server = false;
+ int version_major = 0;
+ int version_minor = 0;
+ std::string port = "19000";
+
+ std::regex metal_version_re{"^-?-std=macos-metal([0-9]+)\\.([0-9]+)"};
+
+ std::vector<std::string> args;
+ for (int i = 1; i < argc; i++) {
+ std::string arg = argv[i];
+ if (arg == "-s" || arg == "--server") {
+ run_server = true;
+ continue;
+ }
+ if (arg == "-p" || arg == "--port") {
+ if (i < argc - 1) {
+ i++;
+ port = argv[i];
+ } else {
+ printf("expected port number");
+ exit(1);
+ }
+ continue;
+ }
+
+ // xcrun flags are ignored so this executable can be used as a replacement for xcrun.
+ if ((arg == "-x" || arg == "-sdk") && (i < argc - 1)) {
+ i++;
+ continue;
+ }
+ if (arg == "metal") {
+ for (; i < argc; i++) {
+ arg = argv[i];
+ // metal_version_re
+ std::smatch metal_version_match;
+ if (std::regex_match(arg, metal_version_match, metal_version_re)) {
+ version_major = std::atoi(metal_version_match[1].str().c_str());
+ version_minor = std::atoi(metal_version_match[2].str().c_str());
+ continue;
+ }
+ if (arg == "-c") {
+ break;
+ }
+ }
+ continue;
+ }
+
+ args.emplace_back(arg);
+ }
+
+ bool success = false;
+
+ if (run_server) {
+ success = RunServer(port);
+ } else {
+ std::string address;
+ std::string file;
+ switch (args.size()) {
+ case 1:
+ TINT_BEGIN_DISABLE_WARNING(DEPRECATED);
+ if (auto* addr = getenv("TINT_REMOTE_COMPILE_ADDRESS")) {
+ address = addr;
+ }
+ TINT_END_DISABLE_WARNING(DEPRECATED);
+ file = args[0];
+ break;
+ case 2:
+ address = args[0];
+ file = args[1];
+ break;
+ default:
+ std::cerr << "expected 1 or 2 arguments, got " << args.size() << std::endl
+ << std::endl;
+ ShowUsage();
+ }
+ if (address.empty() || file.empty()) {
+ ShowUsage();
+ }
+ success = RunClient(address, port, file, version_major, version_minor);
+ }
+
+ if (!success) {
+ exit(1);
+ }
+
+ return 0;
+}
+
+bool RunServer(std::string port) {
+ auto server_socket = Socket::Listen("", port.c_str());
+ if (!server_socket) {
+ std::cout << "Failed to listen on port " << port << std::endl;
+ return false;
+ }
+ std::cout << "Listening on port " << port.c_str() << "..." << std::endl;
+ while (auto conn = server_socket->Accept()) {
+ std::thread([=] {
+ DEBUG("Client connected...");
+ Stream stream{conn.get(), ""};
+
+ {
+ ConnectionRequest req;
+ stream >> req;
+ if (!stream.error.empty()) {
+ DEBUG("%s", stream.error.c_str());
+ return;
+ }
+ ConnectionResponse resp;
+ if (req.protocol_version != kProtocolVersion) {
+ DEBUG("Protocol version mismatch");
+ resp.error = "Protocol version mismatch";
+ stream << resp;
+ return;
+ }
+ stream << resp;
+ }
+ DEBUG("Connection established");
+ {
+ CompileRequest req;
+ stream >> req;
+ if (!stream.error.empty()) {
+ DEBUG("%s\n", stream.error.c_str());
+ return;
+ }
+#ifdef TINT_ENABLE_MSL_VALIDATION_USING_METAL_API
+ if (req.language == SourceLanguage::MSL) {
+ auto version = tint::msl::validate::MslVersion::kMsl_1_2;
+ if (req.version_major == 2 && req.version_minor == 1) {
+ version = tint::msl::validate::MslVersion::kMsl_2_1;
+ }
+ auto result = tint::msl::validate::UsingMetalAPI(req.source, version);
+ CompileResponse resp;
+ if (result.failed) {
+ resp.error = result.output;
+ }
+ stream << resp;
+ return;
+ }
+#endif
+ CompileResponse resp;
+ resp.error = "server cannot compile this type of shader";
+ stream << resp;
+ }
+ }).detach();
+ }
+ return true;
+}
+
+bool RunClient(std::string address,
+ std::string port,
+ std::string file,
+ int version_major,
+ int version_minor) {
+ // Read the file
+ std::ifstream input(file, std::ios::binary);
+ if (!input) {
+ std::cerr << "Couldn't open '" << file << "'" << std::endl;
+ return false;
+ }
+ std::string source((std::istreambuf_iterator<char>(input)), std::istreambuf_iterator<char>());
+
+ constexpr const int timeout_ms = 10000;
+ DEBUG("Connecting to %s:%s...", address.c_str(), port.c_str());
+ auto conn = Socket::Connect(address.c_str(), port.c_str(), timeout_ms);
+ if (!conn) {
+ std::cerr << "Connection failed" << std::endl;
+ return false;
+ }
+
+ Stream stream{conn.get(), ""};
+
+ DEBUG("Sending connection request...");
+ auto conn_resp = Send(stream, ConnectionRequest{kProtocolVersion});
+ if (!stream.error.empty()) {
+ std::cerr << stream.error << std::endl;
+ return false;
+ }
+ if (!conn_resp.error.empty()) {
+ std::cerr << conn_resp.error << std::endl;
+ return false;
+ }
+ DEBUG("Connection established. Requesting compile...");
+ auto comp_resp =
+ Send(stream, CompileRequest{SourceLanguage::MSL, version_major, version_minor, source});
+ if (!stream.error.empty()) {
+ std::cerr << stream.error << std::endl;
+ return false;
+ }
+ if (!comp_resp.error.empty()) {
+ std::cerr << comp_resp.error << std::endl;
+ return false;
+ }
+ DEBUG("Compilation successful");
+ return true;
+}
diff --git a/src/tint/lang/glsl/writer/ast_printer/ast_printer.cc b/src/tint/lang/glsl/writer/ast_printer/ast_printer.cc
index 127a2da..895fd56 100644
--- a/src/tint/lang/glsl/writer/ast_printer/ast_printer.cc
+++ b/src/tint/lang/glsl/writer/ast_printer/ast_printer.cc
@@ -62,6 +62,7 @@
#include "src/tint/lang/wgsl/ast/transform/single_entry_point.h"
#include "src/tint/lang/wgsl/ast/transform/std140.h"
#include "src/tint/lang/wgsl/ast/transform/texture_1d_to_2d.h"
+#include "src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h"
#include "src/tint/lang/wgsl/ast/transform/unshadow.h"
#include "src/tint/lang/wgsl/ast/transform/zero_init_workgroup_memory.h"
#include "src/tint/lang/wgsl/ast/variable_decl_statement.h"
@@ -229,6 +230,15 @@
manager.Add<ast::transform::RemovePhonies>();
+ // TextureBuiltinsFromUniform must come before CombineSamplers to preserve texture binding point
+ // info, instead of combined sampler binding point. As a result, TextureBuiltinsFromUniform also
+ // comes before BindingRemapper so the binding point info it reflects is before remapping.
+ if (options.texture_builtins_from_uniform) {
+ manager.Add<ast::transform::TextureBuiltinsFromUniform>();
+ data.Add<ast::transform::TextureBuiltinsFromUniform::Config>(
+ options.texture_builtins_from_uniform->ubo_binding);
+ }
+
data.Add<ast::transform::CombineSamplers::BindingInfo>(options.binding_map,
options.placeholder_binding_point);
manager.Add<ast::transform::CombineSamplers>();
@@ -254,6 +264,10 @@
SanitizedResult result;
ast::transform::DataMap outputs;
result.program = manager.Run(in, data, outputs);
+ if (auto* res = outputs.Get<ast::transform::TextureBuiltinsFromUniform::Result>()) {
+ result.needs_internal_uniform_buffer = true;
+ result.bindpoint_to_data = std::move(res->bindpoint_to_data);
+ }
return result;
}
diff --git a/src/tint/lang/glsl/writer/ast_printer/ast_printer.h b/src/tint/lang/glsl/writer/ast_printer/ast_printer.h
index 6c8f5fb..ad44120 100644
--- a/src/tint/lang/glsl/writer/ast_printer/ast_printer.h
+++ b/src/tint/lang/glsl/writer/ast_printer/ast_printer.h
@@ -21,6 +21,7 @@
#include <unordered_set>
#include <utility>
+#include "src/tint/api/options/texture_builtins_from_uniform.h"
#include "src/tint/lang/core/builtin_value.h"
#include "src/tint/lang/glsl/writer/common/version.h"
#include "src/tint/lang/wgsl/program/program_builder.h"
@@ -53,6 +54,13 @@
/// The sanitized program.
Program program;
+
+ /// True if the shader needs a UBO.
+ bool needs_internal_uniform_buffer = false;
+
+ /// Store a map of global texture variable binding point to the byte offset and data type to
+ /// push into the internal uniform buffer.
+ TextureBuiltinsFromUniformOptions::BindingPointToFieldAndOffset bindpoint_to_data;
};
/// Sanitize a program in preparation for generating GLSL.
diff --git a/src/tint/lang/glsl/writer/common/options.h b/src/tint/lang/glsl/writer/common/options.h
index e2aadd7..2f55c4b 100644
--- a/src/tint/lang/glsl/writer/common/options.h
+++ b/src/tint/lang/glsl/writer/common/options.h
@@ -15,10 +15,12 @@
#ifndef SRC_TINT_LANG_GLSL_WRITER_COMMON_OPTIONS_H_
#define SRC_TINT_LANG_GLSL_WRITER_COMMON_OPTIONS_H_
+#include <optional>
#include <string>
#include <unordered_map>
#include "src/tint/api/options/external_texture.h"
+#include "src/tint/api/options/texture_builtins_from_uniform.h"
#include "src/tint/lang/core/access.h"
#include "src/tint/lang/glsl/writer/common/version.h"
#include "src/tint/lang/wgsl/sem/sampler_texture_pair.h"
@@ -66,6 +68,11 @@
/// Options used in the binding mappings for external textures
ExternalTextureOptions external_texture_options = {};
+ /// Options used to map WGSL textureNumLevels/textureNumSamples builtins to internal uniform
+ /// buffer values. If not specified, emits corresponding GLSL builtins
+ /// textureQueryLevels/textureSamples directly.
+ std::optional<TextureBuiltinsFromUniformOptions> texture_builtins_from_uniform = std::nullopt;
+
/// The GLSL version to emit
Version version;
@@ -74,6 +81,7 @@
allow_collisions,
disable_workgroup_init,
external_texture_options,
+ texture_builtins_from_uniform,
version);
};
diff --git a/src/tint/lang/glsl/writer/output.h b/src/tint/lang/glsl/writer/output.h
index 084b0ef..286af3f 100644
--- a/src/tint/lang/glsl/writer/output.h
+++ b/src/tint/lang/glsl/writer/output.h
@@ -19,6 +19,7 @@
#include <utility>
#include <vector>
+#include "src/tint/api/options/texture_builtins_from_uniform.h"
#include "src/tint/lang/wgsl/ast/pipeline_stage.h"
namespace tint::glsl::writer {
@@ -39,6 +40,13 @@
/// The list of entry points in the generated GLSL.
std::vector<std::pair<std::string, ast::PipelineStage>> entry_points;
+
+ /// True if the shader needs a UBO.
+ bool needs_internal_uniform_buffer = false;
+
+ /// Store a map of global texture variable binding points to the byte offset and data type to
+ /// push into the internal uniform buffer.
+ TextureBuiltinsFromUniformOptions::BindingPointToFieldAndOffset bindpoint_to_data;
};
} // namespace tint::glsl::writer
diff --git a/src/tint/lang/glsl/writer/writer.cc b/src/tint/lang/glsl/writer/writer.cc
index 4124346..e02c6bb 100644
--- a/src/tint/lang/glsl/writer/writer.cc
+++ b/src/tint/lang/glsl/writer/writer.cc
@@ -15,6 +15,7 @@
#include "src/tint/lang/glsl/writer/writer.h"
#include <memory>
+#include <utility>
#include "src/tint/lang/glsl/writer/ast_printer/ast_printer.h"
#include "src/tint/lang/wgsl/ast/transform/binding_remapper.h"
@@ -43,6 +44,8 @@
Output output;
output.glsl = impl->Result();
+ output.needs_internal_uniform_buffer = sanitized_result.needs_internal_uniform_buffer;
+ output.bindpoint_to_data = std::move(sanitized_result.bindpoint_to_data);
// Collect the list of entry points in the sanitized program.
for (auto* func : sanitized_result.program.AST().Functions()) {
diff --git a/src/tint/lang/msl/validate/val.h b/src/tint/lang/msl/validate/val.h
index 498516b..78096dc 100644
--- a/src/tint/lang/msl/validate/val.h
+++ b/src/tint/lang/msl/validate/val.h
@@ -40,7 +40,7 @@
struct Result {
/// True if validation passed
bool failed = false;
- /// Output of DXC.
+ /// Output of Metal compiler.
std::string output;
};
diff --git a/src/tint/lang/wgsl/ast/transform/BUILD.cmake b/src/tint/lang/wgsl/ast/transform/BUILD.cmake
index e56ea18..d91332e 100644
--- a/src/tint/lang/wgsl/ast/transform/BUILD.cmake
+++ b/src/tint/lang/wgsl/ast/transform/BUILD.cmake
@@ -115,6 +115,8 @@
lang/wgsl/ast/transform/substitute_override.h
lang/wgsl/ast/transform/texture_1d_to_2d.cc
lang/wgsl/ast/transform/texture_1d_to_2d.h
+ lang/wgsl/ast/transform/texture_builtins_from_uniform.cc
+ lang/wgsl/ast/transform/texture_builtins_from_uniform.h
lang/wgsl/ast/transform/transform.cc
lang/wgsl/ast/transform/transform.h
lang/wgsl/ast/transform/truncate_interstage_variables.cc
@@ -213,6 +215,7 @@
lang/wgsl/ast/transform/std140_test.cc
lang/wgsl/ast/transform/substitute_override_test.cc
lang/wgsl/ast/transform/texture_1d_to_2d_test.cc
+ lang/wgsl/ast/transform/texture_builtins_from_uniform_test.cc
lang/wgsl/ast/transform/transform_test.cc
lang/wgsl/ast/transform/truncate_interstage_variables_test.cc
lang/wgsl/ast/transform/unshadow_test.cc
diff --git a/src/tint/lang/wgsl/ast/transform/BUILD.gn b/src/tint/lang/wgsl/ast/transform/BUILD.gn
index 8df2f94..c55205d 100644
--- a/src/tint/lang/wgsl/ast/transform/BUILD.gn
+++ b/src/tint/lang/wgsl/ast/transform/BUILD.gn
@@ -120,6 +120,8 @@
"substitute_override.h",
"texture_1d_to_2d.cc",
"texture_1d_to_2d.h",
+ "texture_builtins_from_uniform.cc",
+ "texture_builtins_from_uniform.h",
"transform.cc",
"transform.h",
"truncate_interstage_variables.cc",
@@ -216,6 +218,7 @@
"std140_test.cc",
"substitute_override_test.cc",
"texture_1d_to_2d_test.cc",
+ "texture_builtins_from_uniform_test.cc",
"transform_test.cc",
"truncate_interstage_variables_test.cc",
"unshadow_test.cc",
diff --git a/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.cc b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.cc
new file mode 100644
index 0000000..4c4baee
--- /dev/null
+++ b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.cc
@@ -0,0 +1,491 @@
+// Copyright 2023 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/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h"
+
+#include <memory>
+#include <queue>
+#include <string>
+#include <utility>
+#include <variant>
+#include <vector>
+
+#include "src/tint/lang/wgsl/program/clone_context.h"
+#include "src/tint/lang/wgsl/program/program_builder.h"
+#include "src/tint/lang/wgsl/resolver/resolve.h"
+#include "src/tint/lang/wgsl/sem/call.h"
+#include "src/tint/lang/wgsl/sem/function.h"
+#include "src/tint/lang/wgsl/sem/module.h"
+#include "src/tint/lang/wgsl/sem/statement.h"
+#include "src/tint/lang/wgsl/sem/variable.h"
+
+#include "src/tint/utils/containers/hashmap.h"
+#include "src/tint/utils/containers/vector.h"
+#include "src/tint/utils/rtti/switch.h"
+
+TINT_INSTANTIATE_TYPEINFO(tint::ast::transform::TextureBuiltinsFromUniform);
+TINT_INSTANTIATE_TYPEINFO(tint::ast::transform::TextureBuiltinsFromUniform::Config);
+TINT_INSTANTIATE_TYPEINFO(tint::ast::transform::TextureBuiltinsFromUniform::Result);
+
+namespace tint::ast::transform {
+
+namespace {
+
+/// The member name of the texture builtin values.
+constexpr std::string_view kTextureBuiltinValuesMemberNamePrefix = "texture_builtin_value_";
+
+bool ShouldRun(const Program* program) {
+ for (auto* fn : program->AST().Functions()) {
+ if (auto* sem_fn = program->Sem().Get(fn)) {
+ for (auto* builtin : sem_fn->DirectlyCalledBuiltins()) {
+ // GLSL ES has no native support for the counterpart of
+ // textureNumLevels (textureQueryLevels) and textureNumSamples (textureSamples)
+ if (builtin->Type() == core::Function::kTextureNumLevels) {
+ return true;
+ }
+ if (builtin->Type() == core::Function::kTextureNumSamples) {
+ return true;
+ }
+ }
+ }
+ }
+ return false;
+}
+
+} // namespace
+
+TextureBuiltinsFromUniform::TextureBuiltinsFromUniform() = default;
+TextureBuiltinsFromUniform::~TextureBuiltinsFromUniform() = default;
+
+/// PIMPL state for the transform
+struct TextureBuiltinsFromUniform::State {
+ /// Constructor
+ /// @param program the source program
+ /// @param in the input transform data
+ /// @param out the output transform data
+ explicit State(const Program* program, const DataMap& in, DataMap& out)
+ : src(program), inputs(in), outputs(out) {}
+
+ /// Runs the transform
+ /// @returns the new program or SkipTransform if the transform is not required
+ ApplyResult Run() {
+ auto* cfg = inputs.Get<Config>();
+ if (cfg == nullptr) {
+ b.Diagnostics().add_error(
+ diag::System::Transform,
+ "missing transform data for " +
+ std::string(tint::TypeInfo::Of<TextureBuiltinsFromUniform>().name));
+ return resolver::Resolve(b);
+ }
+
+ if (!ShouldRun(ctx.src)) {
+ return SkipTransform;
+ }
+
+ // The dependency order declartions guaranteed that we traverse interested functions in the
+ // following order:
+ // 1. texture builtins
+ // 2. user function directly calls texture builtins
+ // 3. user function calls 2.
+ // 4. user function calls 3.
+ // ...
+ // n. entry point function.
+ for (auto* fn_decl : sem.Module()->DependencyOrderedDeclarations()) {
+ if (auto* fn = sem.Get<sem::Function>(fn_decl)) {
+ for (auto* call : fn->DirectCalls()) {
+ auto* call_expr = call->Declaration();
+
+ tint::Switch(
+ call->Target(),
+ [&](const sem::Builtin* builtin) {
+ if (builtin->Type() != core::Function::kTextureNumLevels &&
+ builtin->Type() != core::Function::kTextureNumSamples) {
+ return;
+ }
+ if (auto* call_stmt =
+ call->Stmt()->Declaration()->As<CallStatement>()) {
+ if (call_stmt->expr == call->Declaration()) {
+ // textureNumLevels() / textureNumSamples() is used as a
+ // statement. The argument expression must be side-effect free,
+ // so just drop the statement.
+ RemoveStatement(ctx, call_stmt);
+ return;
+ }
+ }
+
+ auto* texture_expr = call->Declaration()->args[0];
+ auto* texture_sem = sem.GetVal(texture_expr)->RootIdentifier();
+ TINT_ASSERT(texture_sem);
+
+ TextureBuiltinsFromUniformOptions::Field dataType =
+ GetFieldFromBuiltinFunctionType(builtin->Type());
+
+ tint::Switch(
+ texture_sem,
+ [&](const sem::GlobalVariable* global) {
+ // This texture variable is a global variable.
+ auto binding = GetAndRecordGlobalBinding(global, dataType);
+ // Record the call and binding to be replaced later.
+ builtin_to_replace.Add(call_expr, binding);
+ },
+ [&](const sem::Variable* variable) {
+ // This texture variable is a user function parameter.
+ auto new_param =
+ GetAndRecordFunctionParameter(fn, variable, dataType);
+ // Record the call and new_param to be replaced later.
+ builtin_to_replace.Add(call_expr, new_param);
+ },
+ [&](Default) {
+ TINT_ICE() << "unexpected texture root identifier";
+ });
+ },
+ [&](const sem::Function* user_fn) {
+ auto user_param_to_info = fn_to_data.Find(user_fn);
+ if (!user_param_to_info) {
+ // Uninterested function not calling texture builtins with function
+ // texture param.
+ return;
+ }
+ TINT_ASSERT(call->Arguments().Length() ==
+ user_fn->Declaration()->params.Length());
+ for (size_t i = 0; i < call->Arguments().Length(); i++) {
+ auto param = user_fn->Declaration()->params[i];
+ auto info = user_param_to_info->Get(param);
+ if (info.has_value()) {
+ auto* arg = call->Arguments()[i];
+ auto* texture_sem = arg->RootIdentifier();
+ auto& args = call_to_data.GetOrCreate(call_expr, [&] {
+ return Vector<
+ std::variant<BindingPoint, const ast::Parameter*>, 4>();
+ });
+
+ tint::Switch(
+ texture_sem,
+ [&](const sem::GlobalVariable* global) {
+ // This texture variable is a global variable.
+ auto binding =
+ GetAndRecordGlobalBinding(global, info->field);
+ // Record the binding to add to args.
+ args.Push(binding);
+ },
+ [&](const sem::Variable* variable) {
+ // This texture variable is a user function parameter.
+ auto new_param = GetAndRecordFunctionParameter(
+ fn, variable, info->field);
+ // Record adding extra function parameter
+ args.Push(new_param);
+ },
+ [&](Default) {
+ TINT_ICE() << "unexpected texture root identifier";
+ });
+ }
+ }
+ });
+ }
+ }
+ }
+
+ // If there's no interested texture builtin at all, skip the transform.
+ if (bindpoint_to_data.empty()) {
+ return SkipTransform;
+ }
+
+ // If any functions need extra params, add them now.
+ if (!fn_to_data.IsEmpty()) {
+ for (auto pair : fn_to_data) {
+ auto* fn = pair.key;
+
+ // Reorder the param to a vector to make sure params are in the correct order.
+ Vector<const ast::Parameter*, 4> extra_params_in_order;
+ extra_params_in_order.Resize(pair.value.Count());
+ for (auto t_p : pair.value) {
+ TINT_ASSERT(t_p.value.extra_idx < extra_params_in_order.Length());
+ extra_params_in_order[t_p.value.extra_idx] = t_p.value.param;
+ }
+
+ for (auto p : extra_params_in_order) {
+ ctx.InsertBack(fn->Declaration()->params, p);
+ }
+ }
+ }
+
+ // Replace all interested texture builtin calls.
+ for (auto pair : builtin_to_replace) {
+ auto call = pair.key;
+ if (std::holds_alternative<BindingPoint>(pair.value)) {
+ // This texture is a global variable with binding point.
+ // Read builtin value from uniform buffer.
+ auto* builtin_value = GetUniformValue(std::get<BindingPoint>(pair.value));
+ ctx.Replace(call, builtin_value);
+ } else {
+ // Otherwise this value comes from a function param
+ auto* param = std::get<const ast::Parameter*>(pair.value);
+ ctx.Replace(call, b.Expr(param));
+ }
+ }
+
+ // Insert all extra args to interested function calls.
+ for (auto pair : call_to_data) {
+ auto call = pair.key;
+ for (auto new_arg_info : pair.value) {
+ if (std::holds_alternative<BindingPoint>(new_arg_info)) {
+ // This texture is a global variable with binding point.
+ // Read builtin value from uniform buffer.
+ auto* builtin_value = GetUniformValue(std::get<BindingPoint>(new_arg_info));
+ ctx.InsertBack(call->args, builtin_value);
+ } else {
+ // Otherwise this value comes from a function param
+ auto* param = std::get<const ast::Parameter*>(new_arg_info);
+ ctx.InsertBack(call->args, b.Expr(param));
+ }
+ }
+ }
+
+ outputs.Add<Result>(bindpoint_to_data);
+
+ ctx.Clone();
+ return resolver::Resolve(b);
+ }
+
+ private:
+ /// The source program
+ const Program* const src;
+ /// The transform inputs
+ const DataMap& inputs;
+ /// The transform outputs
+ DataMap& outputs;
+ /// The target program builder
+ ProgramBuilder b;
+ /// The clone context
+ program::CloneContext ctx = {&b, src, /* auto_clone_symbols */ true};
+ /// Alias to the semantic info in ctx.src
+ const sem::Info& sem = ctx.src->Sem();
+
+ /// The bindpoint to byte offset and field to pass out in transform result.
+ /// For one texture type, it could only be passed into one of the
+ /// textureNumLevels or textureNumSamples because their accepting param texture
+ /// type is different. There cannot be a binding entry with both field type.
+ /// Note: because this transform must be run before CombineSampler and BindingRemapper,
+ /// the binding number here is before remapped.
+ Result::BindingPointToFieldAndOffset bindpoint_to_data;
+
+ struct FunctionExtraParamInfo {
+ using Field = TextureBuiltinsFromUniformOptions::Field;
+ // The kind of texture information this parameter holds.
+ Field field = Field::TextureNumLevels;
+
+ // The extra passed in param that corresponds to the texture param.
+ const ast::Parameter* param = nullptr;
+
+ // id of this extra param e.g. f(t0, foo, t1, e0, e1) e0 and e1 are extra params, their
+ // extra_idx are 0 and 1. This is to help sort extra ids in the correct order.
+ size_t extra_idx = 0;
+ };
+
+ /// Store a map from function to a collection of extra params that need adding.
+ /// The value of the map is made a map instead of a vector to make it easier to find the param.
+ /// for call sites. e.g. fn f(t: texture_2d<f32>) -> u32 {
+ /// return textureNumLevels(t);
+ /// }
+ /// ->
+ /// fn f(t : texture_2d<f32>, tint_symbol : u32) -> u32 {
+ /// return tint_symbol;
+ /// }
+ Hashmap<const sem::Function*, Hashmap<const ast::Parameter*, FunctionExtraParamInfo, 4>, 8>
+ fn_to_data;
+
+ /// For each callsite of the above functions, record a vector of extra call args that need
+ /// inserting. e.g. f(tex)
+ /// ->
+ /// f(tex, internal_uniform.texture_builtin_value), if tex is from a global
+ /// variable, store the BindingPoint. or f(tex, extra_param_tex), if tex is from a function
+ /// param, store the texture function parameter pointer.
+ Hashmap<const CallExpression*, Vector<std::variant<BindingPoint, const ast::Parameter*>, 4>, 8>
+ call_to_data;
+
+ /// Texture builtin calls to be replaced by either uniform values or function parameters.
+ Hashmap<const CallExpression*, std::variant<BindingPoint, const ast::Parameter*>, 8>
+ builtin_to_replace;
+
+ /// A map from global texture bindpoint to the symbol storing its builtin value in the uniform
+ /// buffer struct.
+ Hashmap<BindingPoint, Symbol, 16> bindpoint_to_syms;
+
+ /// The internal uniform buffer
+ const Variable* ubo = nullptr;
+ /// Get or create a UBO including u32 scalars for texture builtin values.
+ /// @returns the symbol of the uniform buffer variable.
+ Symbol GetUboSym() {
+ if (ubo) {
+ // Already created
+ return ubo->name->symbol;
+ }
+
+ auto* cfg = inputs.Get<Config>();
+
+ Vector<const ast::StructMember*, 16> new_members;
+ new_members.Resize(bindpoint_to_data.size());
+ for (auto it : bindpoint_to_data) {
+ // Emit a u32 scalar for each binding that needs builtin value passed in.
+ size_t i = it.second.second / sizeof(uint32_t);
+ TINT_ASSERT(i < new_members.Length());
+ // Append the vector index with the variable name to avoid unstable naming issue.
+ auto sym = b.Symbols().New(std::string(kTextureBuiltinValuesMemberNamePrefix) +
+ std::to_string(i));
+ bindpoint_to_syms.Add(it.first, sym);
+ new_members[i] = b.Member(sym, b.ty.u32());
+ }
+
+ // Find if there's any existing global variable using the same cfg->ubo_binding
+ for (auto* var : src->AST().Globals<Var>()) {
+ if (var->HasBindingPoint()) {
+ auto* global_sem = sem.Get<sem::GlobalVariable>(var);
+
+ // The original binding point
+ BindingPoint binding_point = *global_sem->BindingPoint();
+
+ if (binding_point == cfg->ubo_binding) {
+ // This ubo_binding struct already exists.
+ // which should only be added by other *FromUniform transforms.
+ // Replace it with a new struct including the new_member.
+ // Then remove the old structure global declaration.
+
+ ubo = var->As<Variable>();
+
+ auto* ty = global_sem->Type()->UnwrapRef();
+ auto* str = ty->As<sem::Struct>();
+ if (TINT_UNLIKELY(!str)) {
+ TINT_ICE()
+ << "existing ubo binding " << cfg->ubo_binding << " is not a struct.";
+ return ctx.Clone(ubo->name->symbol);
+ }
+
+ for (auto new_member : new_members) {
+ ctx.InsertBack(str->Declaration()->members, new_member);
+ }
+ return ctx.Clone(ubo->name->symbol);
+ }
+ }
+ }
+
+ auto* buffer_struct = b.Structure(b.Sym(), std::move(new_members));
+ ubo = b.GlobalVar(b.Sym(), b.ty.Of(buffer_struct), core::AddressSpace::kUniform,
+ b.Group(core::AInt(cfg->ubo_binding.group)),
+ b.Binding(core::AInt(cfg->ubo_binding.binding)));
+ return ubo->name->symbol;
+ }
+
+ /// Get the expression of retrieving the builtin value from the uniform buffer.
+ /// @param binding of the global variable.
+ /// @returns an expression of the builtin value.
+ const ast::Expression* GetUniformValue(const BindingPoint& binding) {
+ auto iter = bindpoint_to_data.find(binding);
+ TINT_ASSERT(iter != bindpoint_to_data.end());
+
+ // Make sure GetUboSym() is called first to initialize the uniform buffer struct.
+ auto ubo_sym = GetUboSym();
+ // Load the builtin value from the UBO.
+ auto member_sym = bindpoint_to_syms.Get(binding);
+ TINT_ASSERT(member_sym.has_value());
+ auto* builtin_value = b.MemberAccessor(ubo_sym, *member_sym);
+
+ return builtin_value;
+ }
+
+ /// Get and return the binding of the global texture variable. Record in bindpoint_to_data if
+ /// first visited.
+ /// @param global global variable of the texture variable.
+ /// @param field type of the interested builtin function data related to this texture.
+ /// @returns binding of the global variable.
+ BindingPoint GetAndRecordGlobalBinding(const sem::GlobalVariable* global,
+ TextureBuiltinsFromUniformOptions::Field field) {
+ auto binding = global->BindingPoint().value();
+ auto iter = bindpoint_to_data.find(binding);
+ if (iter == bindpoint_to_data.end()) {
+ // First visit, recording the binding.
+ uint32_t index = static_cast<uint32_t>(bindpoint_to_data.size());
+ bindpoint_to_data.emplace(
+ binding,
+ Result::FieldAndOffset{field, index * static_cast<uint32_t>(sizeof(uint32_t))});
+ }
+ return binding;
+ }
+
+ /// Find which function param is the given texture variable.
+ /// Add a new u32 param relates to this texture param. Record in fn_to_data if first visited.
+ /// @param fn the current function scope.
+ /// @param var the texture variable.
+ /// @param field type of the interested builtin function data related to this texture.
+ /// @returns the new u32 function parameter.
+ const ast::Parameter* GetAndRecordFunctionParameter(
+ const sem::Function* fn,
+ const sem::Variable* var,
+ TextureBuiltinsFromUniformOptions::Field field) {
+ auto& param_to_info = fn_to_data.GetOrCreate(
+ fn, [&] { return Hashmap<const ast::Parameter*, FunctionExtraParamInfo, 4>(); });
+
+ const ast::Parameter* param = nullptr;
+ for (auto p : fn->Declaration()->params) {
+ if (p->As<Variable>() == var->Declaration()) {
+ param = p;
+ break;
+ }
+ }
+ TINT_ASSERT(param);
+ // Get or record a new u32 param to this function if first visited.
+ auto entry = param_to_info.Get(param);
+ if (entry.has_value()) {
+ return entry->param;
+ }
+ const ast::Parameter* new_param = b.Param(b.Sym(), b.ty.u32());
+ size_t idx = param_to_info.Count();
+ param_to_info.Add(param, FunctionExtraParamInfo{field, new_param, idx});
+ return new_param;
+ }
+
+ /// Get the uniform options field for the builtin function.
+ /// @param type of the builtin function
+ /// @returns corresponding TextureBuiltinsFromUniformOptions::Field for the builtin
+ static TextureBuiltinsFromUniformOptions::Field GetFieldFromBuiltinFunctionType(
+ core::Function type) {
+ switch (type) {
+ case core::Function::kTextureNumLevels:
+ return TextureBuiltinsFromUniformOptions::Field::TextureNumLevels;
+ case core::Function::kTextureNumSamples:
+ return TextureBuiltinsFromUniformOptions::Field::TextureNumSamples;
+ default:
+ TINT_UNREACHABLE() << "unsupported builtin function type " << type;
+ }
+ return TextureBuiltinsFromUniformOptions::Field::TextureNumLevels;
+ }
+};
+
+Transform::ApplyResult TextureBuiltinsFromUniform::Apply(const Program* src,
+ const DataMap& inputs,
+ DataMap& outputs) const {
+ return State{src, inputs, outputs}.Run();
+}
+
+TextureBuiltinsFromUniform::Config::Config(BindingPoint ubo_bp) : ubo_binding(ubo_bp) {}
+TextureBuiltinsFromUniform::Config::Config(const Config&) = default;
+TextureBuiltinsFromUniform::Config& TextureBuiltinsFromUniform::Config::operator=(const Config&) =
+ default;
+TextureBuiltinsFromUniform::Config::~Config() = default;
+
+TextureBuiltinsFromUniform::Result::Result(BindingPointToFieldAndOffset bindpoint_to_data_in)
+ : bindpoint_to_data(std::move(bindpoint_to_data_in)) {}
+TextureBuiltinsFromUniform::Result::Result(const Result&) = default;
+TextureBuiltinsFromUniform::Result::~Result() = default;
+
+} // namespace tint::ast::transform
diff --git a/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h
new file mode 100644
index 0000000..0db043e
--- /dev/null
+++ b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h
@@ -0,0 +1,119 @@
+// Copyright 2023 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_TINT_LANG_WGSL_AST_TRANSFORM_TEXTURE_BUILTINS_FROM_UNIFORM_H_
+#define SRC_TINT_LANG_WGSL_AST_TRANSFORM_TEXTURE_BUILTINS_FROM_UNIFORM_H_
+
+#include <unordered_map>
+#include <unordered_set>
+
+#include "src/tint/api/common/binding_point.h"
+#include "src/tint/api/options/texture_builtins_from_uniform.h"
+#include "src/tint/lang/wgsl/ast/transform/transform.h"
+
+// Forward declarations
+namespace tint {
+class CloneContext;
+} // namespace tint
+
+namespace tint::ast::transform {
+
+/// TextureBuiltinsFromUniform is a transform that implements calls to textureNumLevels() and
+/// textureNumSamples() by retrieving the texture information from a uniform buffer, as those
+/// builtin functions are not available in some version of GLSL.
+///
+/// The generated uniform buffer will have the form:
+/// ```
+/// struct internal_uniform {
+/// texture_builtin_value_0 : u32,
+/// };
+///
+/// @group(0) @binding(0) var tex : texture_2d<f32>;
+/// ```
+/// The binding group and number used for this uniform buffer are provided via
+/// the `Config` transform input.
+///
+/// The transform coverts the texture builtins calls into values lookup from the internal
+/// buffer. If the texture is a function parameter instead of a global variable, this transform
+/// also takes care of adding extra paramters and arguments to these functions and their callsites.
+///
+/// This transform must run before `CombineSamplers` transform so that the binding point of the
+/// original texture object can be preserved.
+class TextureBuiltinsFromUniform final : public Castable<TextureBuiltinsFromUniform, Transform> {
+ public:
+ /// Constructor
+ TextureBuiltinsFromUniform();
+ /// Destructor
+ ~TextureBuiltinsFromUniform() override;
+
+ /// Configuration options for the TextureBuiltinsFromUniform transform.
+ struct Config final : public Castable<Config, Data> {
+ /// Constructor
+ /// @param ubo_bp the binding point to use for the generated uniform buffer.
+ explicit Config(BindingPoint ubo_bp);
+
+ /// Copy constructor
+ Config(const Config&);
+
+ /// Copy assignment
+ /// @return this Config
+ Config& operator=(const Config&);
+
+ /// Destructor
+ ~Config() override;
+
+ /// The binding point to use for the generated uniform buffer.
+ BindingPoint ubo_binding;
+ };
+
+ /// Information produced about what the transform did.
+ /// If there were no calls to the textureNumLevels() or textureNumSamples() builtin, then no
+ /// Result will be emitted.
+ struct Result final : public Castable<Result, Data> {
+ /// Using for shorter names
+ /// Records the field and the byte offset of the data to push in the internal uniform
+ /// buffer.
+ using FieldAndOffset = TextureBuiltinsFromUniformOptions::FieldAndOffset;
+ /// Maps from binding point to data entry with the information to populate the data.
+ using BindingPointToFieldAndOffset =
+ TextureBuiltinsFromUniformOptions::BindingPointToFieldAndOffset;
+
+ /// Constructor
+ /// @param bindpoint_to_data_in mapping from binding points of global texture variables to
+ /// the byte offsets and data types needed to be pushed into the internal uniform buffer.
+ explicit Result(BindingPointToFieldAndOffset bindpoint_to_data_in);
+
+ /// Copy constructor
+ Result(const Result&);
+
+ /// Destructor
+ ~Result() override;
+
+ /// A map of global texture variable binding point to the byte offset and data type to push
+ /// into the internal uniform buffer.
+ BindingPointToFieldAndOffset bindpoint_to_data;
+ };
+
+ /// @copydoc Transform::Apply
+ ApplyResult Apply(const Program* program,
+ const DataMap& inputs,
+ DataMap& outputs) const override;
+
+ private:
+ struct State;
+};
+
+} // namespace tint::ast::transform
+
+#endif // SRC_TINT_LANG_WGSL_AST_TRANSFORM_TEXTURE_BUILTINS_FROM_UNIFORM_H_
diff --git a/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform_test.cc b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform_test.cc
new file mode 100644
index 0000000..21185e8
--- /dev/null
+++ b/src/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform_test.cc
@@ -0,0 +1,687 @@
+// Copyright 2023 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/tint/lang/wgsl/ast/transform/texture_builtins_from_uniform.h"
+
+#include <utility>
+
+#include "src/tint/lang/wgsl/ast/transform/helper_test.h"
+
+namespace tint::ast::transform {
+namespace {
+
+using TextureBuiltinsFromUniformTest = TransformTest;
+
+TEST_F(TextureBuiltinsFromUniformTest, ShouldRunEmptyModule) {
+ auto* src = R"()";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ EXPECT_FALSE(ShouldRun<TextureBuiltinsFromUniform>(src, data));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, ShouldRunNoTextureNumLevels) {
+ auto* src = R"(
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ _ = textureDimensions(t);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ EXPECT_FALSE(ShouldRun<TextureBuiltinsFromUniform>(src, data));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, ShouldRunWithTextureNumLevels) {
+ auto* src = R"(
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumLevels(t);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ EXPECT_TRUE(ShouldRun<TextureBuiltinsFromUniform>(src, data));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, Error_MissingTransformData) {
+ auto* src = R"(
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumLevels(t);
+}
+)";
+
+ auto* expect =
+ "error: missing transform data for tint::ast::transform::TextureBuiltinsFromUniform";
+
+ auto got = Run<TextureBuiltinsFromUniform>(src);
+
+ EXPECT_EQ(expect, str(got));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, BasicTextureNumLevels) {
+ auto* src = R"(
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumLevels(t);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = tint_symbol_1.texture_builtin_value_0;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ // Note: Using the following EXPECT_EQ directly on BindingPointToFieldAndOffset seems to cause
+ // compiler to hang. EXPECT_EQ(
+ // TextureBuiltinsFromUniformOptions::BindingPointToFieldAndOffset{
+ // {BindgPoint{0u, 0u},
+ // std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u)}},
+ // val->bindpoint_to_data);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, BasicTextureNumSamples) {
+ auto* src = R"(
+@group(0) @binding(0) var t : texture_multisampled_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var samples : u32 = textureNumSamples(t);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var t : texture_multisampled_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var samples : u32 = tint_symbol_1.texture_builtin_value_0;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumSamples, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, SameBuiltinCalledMultipleTimes) {
+ auto* src = R"(
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumLevels(tex);
+ len = textureNumLevels(tex);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = tint_symbol_1.texture_builtin_value_0;
+ len = tint_symbol_1.texture_builtin_value_0;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, SameBuiltinCalledMultipleTimesTextureNumSamples) {
+ auto* src = R"(
+@group(0) @binding(0) var tex : texture_multisampled_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumSamples(tex);
+ len = textureNumSamples(tex);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var tex : texture_multisampled_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = tint_symbol_1.texture_builtin_value_0;
+ len = tint_symbol_1.texture_builtin_value_0;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumSamples, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, TextureAsFunctionParameterBasic) {
+ auto* src = R"(
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f(t: texture_2d<f32>) -> u32 {
+ return textureNumLevels(t);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_1 {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_2 : tint_symbol_1;
+
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f(t : texture_2d<f32>, tint_symbol : u32) -> u32 {
+ return tint_symbol;
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex, tint_symbol_2.texture_builtin_value_0);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, TextureAsFunctionParameterUsedTwice) {
+ auto* src = R"(
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f(t: texture_2d<f32>) -> u32 {
+ var len = textureNumLevels(t);
+ len += textureNumLevels(t);
+ return len;
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_1 {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_2 : tint_symbol_1;
+
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f(t : texture_2d<f32>, tint_symbol : u32) -> u32 {
+ var len = tint_symbol;
+ len += tint_symbol;
+ return len;
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex, tint_symbol_2.texture_builtin_value_0);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, TextureAsFunctionParameterMultipleParameters) {
+ auto* src = R"(
+@group(0) @binding(0) var tex1 : texture_2d<f32>;
+@group(0) @binding(1) var tex2 : texture_2d<f32>;
+@group(0) @binding(2) var tex3 : texture_2d<f32>;
+
+fn f(t1: texture_2d<f32>, t2: texture_2d<f32>, t3: texture_2d<f32>) -> u32 {
+ return textureNumLevels(t1) + textureNumLevels(t2) + textureNumLevels(t3);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex1, tex2, tex3);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_3 {
+ texture_builtin_value_0 : u32,
+ texture_builtin_value_1 : u32,
+ texture_builtin_value_2 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_4 : tint_symbol_3;
+
+@group(0) @binding(0) var tex1 : texture_2d<f32>;
+
+@group(0) @binding(1) var tex2 : texture_2d<f32>;
+
+@group(0) @binding(2) var tex3 : texture_2d<f32>;
+
+fn f(t1 : texture_2d<f32>, t2 : texture_2d<f32>, t3 : texture_2d<f32>, tint_symbol : u32, tint_symbol_1 : u32, tint_symbol_2 : u32) -> u32 {
+ return ((tint_symbol + tint_symbol_1) + tint_symbol_2);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f(tex1, tex2, tex3, tint_symbol_4.texture_builtin_value_0, tint_symbol_4.texture_builtin_value_1, tint_symbol_4.texture_builtin_value_2);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(3u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 4u),
+ val->bindpoint_to_data.at(BindingPoint{0, 1}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 8u),
+ val->bindpoint_to_data.at(BindingPoint{0, 2}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, TextureAsFunctionParameterNested) {
+ auto* src = R"(
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f2(tt: texture_2d<f32>) -> u32 {
+ return textureNumLevels(tt);
+}
+
+fn f1(t: texture_2d<f32>) -> u32 {
+ return f2(t);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f1(tex);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_2 {
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_3 : tint_symbol_2;
+
+@group(0) @binding(0) var tex : texture_2d<f32>;
+
+fn f2(tt : texture_2d<f32>, tint_symbol : u32) -> u32 {
+ return tint_symbol;
+}
+
+fn f1(t : texture_2d<f32>, tint_symbol_1 : u32) -> u32 {
+ return f2(t, tint_symbol_1);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = f1(tex, tint_symbol_3.texture_builtin_value_0);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(1u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, TextureAsFunctionParameterMixed) {
+ auto* src = R"(
+@group(0) @binding(0) var tex0 : texture_2d<f32>;
+@group(0) @binding(1) var tex1 : texture_2d<f32>;
+@group(0) @binding(2) var tex2 : texture_2d<f32>;
+@group(0) @binding(3) var tex3 : texture_2d<f32>;
+@group(0) @binding(4) var tex4 : texture_2d_array<f32>; // unused for textureNumLevels
+
+fn f_nested(t1: texture_2d<f32>, t2: texture_2d<f32>) -> u32 {
+ return textureNumLevels(t1) + textureNumLevels(t2);
+}
+
+fn f1(a: u32, t: texture_2d<f32>) -> u32 {
+ return a + f_nested(t, tex1) + textureNumLevels(tex3);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ _ = textureNumLayers(tex4);
+ _ = f1(9u, tex0);
+ _ = f_nested(tex2, tex2);
+ _ = f_nested(tex1, tex0);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol_3 {
+ texture_builtin_value_0 : u32,
+ texture_builtin_value_1 : u32,
+ texture_builtin_value_2 : u32,
+ texture_builtin_value_3 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_4 : tint_symbol_3;
+
+@group(0) @binding(0) var tex0 : texture_2d<f32>;
+
+@group(0) @binding(1) var tex1 : texture_2d<f32>;
+
+@group(0) @binding(2) var tex2 : texture_2d<f32>;
+
+@group(0) @binding(3) var tex3 : texture_2d<f32>;
+
+@group(0) @binding(4) var tex4 : texture_2d_array<f32>;
+
+fn f_nested(t1 : texture_2d<f32>, t2 : texture_2d<f32>, tint_symbol : u32, tint_symbol_1 : u32) -> u32 {
+ return (tint_symbol + tint_symbol_1);
+}
+
+fn f1(a : u32, t : texture_2d<f32>, tint_symbol_2 : u32) -> u32 {
+ return ((a + f_nested(t, tex1, tint_symbol_2, tint_symbol_4.texture_builtin_value_0)) + tint_symbol_4.texture_builtin_value_1);
+}
+
+@compute @workgroup_size(1)
+fn main() {
+ _ = textureNumLayers(tex4);
+ _ = f1(9u, tex0, tint_symbol_4.texture_builtin_value_2);
+ _ = f_nested(tex2, tex2, tint_symbol_4.texture_builtin_value_3, tint_symbol_4.texture_builtin_value_3);
+ _ = f_nested(tex1, tex0, tint_symbol_4.texture_builtin_value_0, tint_symbol_4.texture_builtin_value_2);
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(4u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 1}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 4u),
+ val->bindpoint_to_data.at(BindingPoint{0, 3}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 8u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 12u),
+ val->bindpoint_to_data.at(BindingPoint{0, 2}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, MultipleTextures) {
+ auto* src = R"(
+@group(0) @binding(0) var t0 : texture_2d<f32>;
+@group(0) @binding(1) var t1 : texture_multisampled_2d<f32>;
+@group(0) @binding(2) var t2 : texture_2d_array<f32>;
+@group(0) @binding(3) var t3 : texture_cube<f32>;
+@group(0) @binding(4) var t4 : texture_depth_2d;
+@group(1) @binding(0) var t5 : texture_depth_multisampled_2d;
+
+@compute @workgroup_size(1)
+fn main() {
+ _ = textureNumLevels(t0);
+ _ = textureNumSamples(t1);
+ _ = textureNumLevels(t2);
+ _ = textureNumLevels(t3);
+ _ = textureNumLevels(t4);
+ _ = textureNumSamples(t5);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ texture_builtin_value_0 : u32,
+ texture_builtin_value_1 : u32,
+ texture_builtin_value_2 : u32,
+ texture_builtin_value_3 : u32,
+ texture_builtin_value_4 : u32,
+ texture_builtin_value_5 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var t0 : texture_2d<f32>;
+
+@group(0) @binding(1) var t1 : texture_multisampled_2d<f32>;
+
+@group(0) @binding(2) var t2 : texture_2d_array<f32>;
+
+@group(0) @binding(3) var t3 : texture_cube<f32>;
+
+@group(0) @binding(4) var t4 : texture_depth_2d;
+
+@group(1) @binding(0) var t5 : texture_depth_multisampled_2d;
+
+@compute @workgroup_size(1)
+fn main() {
+ _ = tint_symbol_1.texture_builtin_value_0;
+ _ = tint_symbol_1.texture_builtin_value_1;
+ _ = tint_symbol_1.texture_builtin_value_2;
+ _ = tint_symbol_1.texture_builtin_value_3;
+ _ = tint_symbol_1.texture_builtin_value_4;
+ _ = tint_symbol_1.texture_builtin_value_5;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(6u, val->bindpoint_to_data.size());
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumSamples, 4u),
+ val->bindpoint_to_data.at(BindingPoint{0, 1}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 8u),
+ val->bindpoint_to_data.at(BindingPoint{0, 2}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 12u),
+ val->bindpoint_to_data.at(BindingPoint{0, 3}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 16u),
+ val->bindpoint_to_data.at(BindingPoint{0, 4}));
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumSamples, 20u),
+ val->bindpoint_to_data.at(BindingPoint{1, 0}));
+}
+
+TEST_F(TextureBuiltinsFromUniformTest, BindingPointExist) {
+ auto* src = R"(
+struct tint_symbol {
+ foo : array<vec4<u32>, 1u>,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = textureNumLevels(t);
+}
+)";
+
+ auto* expect = R"(
+struct tint_symbol {
+ foo : array<vec4<u32>, 1u>,
+ texture_builtin_value_0 : u32,
+}
+
+@group(0) @binding(30) var<uniform> tint_symbol_1 : tint_symbol;
+
+@group(0) @binding(0) var t : texture_2d<f32>;
+
+@compute @workgroup_size(1)
+fn main() {
+ var len : u32 = tint_symbol_1.texture_builtin_value_0;
+}
+)";
+
+ TextureBuiltinsFromUniform::Config cfg({0, 30u});
+
+ DataMap data;
+ data.Add<TextureBuiltinsFromUniform::Config>(std::move(cfg));
+
+ auto got = Run<TextureBuiltinsFromUniform>(src, data);
+
+ EXPECT_EQ(expect, str(got));
+ auto* val = got.data.Get<TextureBuiltinsFromUniform::Result>();
+ ASSERT_NE(val, nullptr);
+ EXPECT_EQ(std::make_pair(TextureBuiltinsFromUniformOptions::Field::TextureNumLevels, 0u),
+ val->bindpoint_to_data.at(BindingPoint{0, 0}));
+}
+
+} // namespace
+} // namespace tint::ast::transform
diff --git a/src/tint/utils/BUILD.cmake b/src/tint/utils/BUILD.cmake
index b4c50f5..0f7a6a9 100644
--- a/src/tint/utils/BUILD.cmake
+++ b/src/tint/utils/BUILD.cmake
@@ -35,6 +35,7 @@
include(utils/reflection/BUILD.cmake)
include(utils/result/BUILD.cmake)
include(utils/rtti/BUILD.cmake)
+include(utils/socket/BUILD.cmake)
include(utils/strconv/BUILD.cmake)
include(utils/symbol/BUILD.cmake)
include(utils/text/BUILD.cmake)
diff --git a/src/tint/utils/macros/compiler.h b/src/tint/utils/macros/compiler.h
index f0af760..6d31ca8 100644
--- a/src/tint/utils/macros/compiler.h
+++ b/src/tint/utils/macros/compiler.h
@@ -31,6 +31,7 @@
#define TINT_DISABLE_WARNING_UNREACHABLE_CODE __pragma(warning(disable : 4702))
#define TINT_DISABLE_WARNING_WEAK_VTABLES /* currently no-op */
#define TINT_DISABLE_WARNING_FLOAT_EQUAL /* currently no-op */
+#define TINT_DISABLE_WARNING_DEPRECATED __pragma(warning(disable : 4996))
// clang-format off
#define TINT_BEGIN_DISABLE_WARNING(name) \
@@ -57,6 +58,7 @@
#define TINT_DISABLE_WARNING_UNREACHABLE_CODE /* currently no-op */
#define TINT_DISABLE_WARNING_WEAK_VTABLES _Pragma("clang diagnostic ignored \"-Wweak-vtables\"")
#define TINT_DISABLE_WARNING_FLOAT_EQUAL _Pragma("clang diagnostic ignored \"-Wfloat-equal\"")
+#define TINT_DISABLE_WARNING_DEPRECATED /* currently no-op */
// clang-format off
#define TINT_BEGIN_DISABLE_WARNING(name) \
@@ -83,6 +85,7 @@
#define TINT_DISABLE_WARNING_UNREACHABLE_CODE /* currently no-op */
#define TINT_DISABLE_WARNING_WEAK_VTABLES /* currently no-op */
#define TINT_DISABLE_WARNING_FLOAT_EQUAL /* currently no-op */
+#define TINT_DISABLE_WARNING_DEPRECATED /* currently no-op */
// clang-format off
#define TINT_BEGIN_DISABLE_WARNING(name) \
diff --git a/src/tint/utils/socket/BUILD.cmake b/src/tint/utils/socket/BUILD.cmake
new file mode 100644
index 0000000..84577fa
--- /dev/null
+++ b/src/tint/utils/socket/BUILD.cmake
@@ -0,0 +1,37 @@
+# Copyright 2023 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.
+
+################################################################################
+# File generated by tools/src/cmd/gen
+# using the template:
+# tools/src/cmd/gen/build/BUILD.cmake.tmpl
+#
+# Do not modify this file directly
+################################################################################
+
+################################################################################
+# Target: tint_utils_socket
+# Kind: lib
+################################################################################
+tint_add_target(tint_utils_socket lib
+ utils/socket/rwmutex.h
+ utils/socket/socket.cc
+ utils/socket/socket.h
+)
+
+if(IS_WIN)
+ tint_target_add_external_dependencies(tint_utils_socket lib
+ "winsock"
+ )
+endif(IS_WIN)
diff --git a/src/tint/utils/socket/BUILD.gn b/src/tint/utils/socket/BUILD.gn
new file mode 100644
index 0000000..7f0bba4
--- /dev/null
+++ b/src/tint/utils/socket/BUILD.gn
@@ -0,0 +1,38 @@
+# Copyright 2023 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.
+
+################################################################################
+# File generated by tools/src/cmd/gen
+# using the template:
+# tools/src/cmd/gen/build/BUILD.gn.tmpl
+#
+# Do not modify this file directly
+################################################################################
+
+import("../../../../scripts/tint_overrides_with_defaults.gni")
+
+import("${tint_src_dir}/tint.gni")
+
+libtint_source_set("socket") {
+ sources = [
+ "rwmutex.h",
+ "socket.cc",
+ "socket.h",
+ ]
+ deps = []
+
+ if (is_win) {
+ deps += [ "${tint_src_dir}:winsock" ]
+ }
+}
diff --git a/src/tint/utils/socket/rwmutex.h b/src/tint/utils/socket/rwmutex.h
new file mode 100644
index 0000000..9b010ed
--- /dev/null
+++ b/src/tint/utils/socket/rwmutex.h
@@ -0,0 +1,190 @@
+// Copyright 2020 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
+//
+// https://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_TINT_UTILS_SOCKET_RWMUTEX_H_
+#define SRC_TINT_UTILS_SOCKET_RWMUTEX_H_
+
+#include <condition_variable>
+#include <mutex>
+
+////////////////////////////////////////////////////////////////////////////////
+// RWMutex
+////////////////////////////////////////////////////////////////////////////////
+
+/// A RWMutex is a reader/writer mutual exclusion lock.
+/// The lock can be held by an arbitrary number of readers or a single writer.
+/// Also known as a shared mutex.
+class RWMutex {
+ public:
+ inline RWMutex() = default;
+
+ /// LockReader() locks the mutex for reading.
+ /// Multiple read locks can be held while there are no writer locks.
+ inline void LockReader();
+
+ /// UnlockReader() unlocks the mutex for reading.
+ inline void UnlockReader();
+
+ /// LockWriter() locks the mutex for writing.
+ /// If the lock is already locked for reading or writing, LockWriter blocks
+ /// until the lock is available.
+ inline void LockWriter();
+
+ /// UnlockWriter() unlocks the mutex for writing.
+ inline void UnlockWriter();
+
+ private:
+ RWMutex(const RWMutex&) = delete;
+ RWMutex& operator=(const RWMutex&) = delete;
+
+ int read_locks = 0;
+ int pending_write_locks = 0;
+ std::mutex mutex;
+ std::condition_variable cv;
+};
+
+void RWMutex::LockReader() {
+ std::unique_lock<std::mutex> lock(mutex);
+ read_locks++;
+}
+
+void RWMutex::UnlockReader() {
+ std::unique_lock<std::mutex> lock(mutex);
+ read_locks--;
+ if (read_locks == 0 && pending_write_locks > 0) {
+ cv.notify_one();
+ }
+}
+
+void RWMutex::LockWriter() {
+ std::unique_lock<std::mutex> lock(mutex);
+ if (read_locks > 0) {
+ pending_write_locks++;
+ cv.wait(lock, [&] { return read_locks == 0; });
+ pending_write_locks--;
+ }
+ lock.release(); // Keep lock held
+}
+
+void RWMutex::UnlockWriter() {
+ if (pending_write_locks > 0) {
+ cv.notify_one();
+ }
+ mutex.unlock();
+}
+
+////////////////////////////////////////////////////////////////////////////////
+// RLock
+////////////////////////////////////////////////////////////////////////////////
+
+/// RLock is a RAII read lock helper for a RWMutex.
+class RLock {
+ public:
+ /// Constructor.
+ /// Locks `mutex` with a read-lock for the lifetime of the WLock.
+ /// @param mutex the mutex
+ explicit inline RLock(RWMutex& mutex);
+ /// Destructor.
+ /// Unlocks the RWMutex.
+ inline ~RLock();
+
+ /// Move constructor
+ /// @param other the other RLock to move into this RLock.
+ inline RLock(RLock&& other);
+ /// Move assignment operator
+ /// @param other the other RLock to move into this RLock.
+ /// @returns this RLock so calls can be chained
+ inline RLock& operator=(RLock&& other);
+
+ private:
+ RLock(const RLock&) = delete;
+ RLock& operator=(const RLock&) = delete;
+
+ RWMutex* m;
+};
+
+RLock::RLock(RWMutex& mutex) : m(&mutex) {
+ m->LockReader();
+}
+
+RLock::~RLock() {
+ if (m != nullptr) {
+ m->UnlockReader();
+ }
+}
+
+RLock::RLock(RLock&& other) {
+ m = other.m;
+ other.m = nullptr;
+}
+
+RLock& RLock::operator=(RLock&& other) {
+ m = other.m;
+ other.m = nullptr;
+ return *this;
+}
+
+////////////////////////////////////////////////////////////////////////////////
+// WLock
+////////////////////////////////////////////////////////////////////////////////
+
+/// WLock is a RAII write lock helper for a RWMutex.
+class WLock {
+ public:
+ /// Constructor.
+ /// Locks `mutex` with a write-lock for the lifetime of the WLock.
+ /// @param mutex the mutex
+ explicit inline WLock(RWMutex& mutex);
+
+ /// Destructor.
+ /// Unlocks the RWMutex.
+ inline ~WLock();
+
+ /// Move constructor
+ /// @param other the other WLock to move into this WLock.
+ inline WLock(WLock&& other);
+ /// Move assignment operator
+ /// @param other the other WLock to move into this WLock.
+ /// @returns this WLock so calls can be chained
+ inline WLock& operator=(WLock&& other);
+
+ private:
+ WLock(const WLock&) = delete;
+ WLock& operator=(const WLock&) = delete;
+
+ RWMutex* m;
+};
+
+WLock::WLock(RWMutex& mutex) : m(&mutex) {
+ m->LockWriter();
+}
+
+WLock::~WLock() {
+ if (m != nullptr) {
+ m->UnlockWriter();
+ }
+}
+
+WLock::WLock(WLock&& other) {
+ m = other.m;
+ other.m = nullptr;
+}
+
+WLock& WLock::operator=(WLock&& other) {
+ m = other.m;
+ other.m = nullptr;
+ return *this;
+}
+
+#endif // SRC_TINT_UTILS_SOCKET_RWMUTEX_H_
diff --git a/src/tint/utils/socket/socket.cc b/src/tint/utils/socket/socket.cc
new file mode 100644
index 0000000..54cb462
--- /dev/null
+++ b/src/tint/utils/socket/socket.cc
@@ -0,0 +1,317 @@
+// 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
+//
+// https://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/tint/utils/socket/socket.h"
+
+#include "src/tint/utils/socket/rwmutex.h"
+
+#if defined(_WIN32)
+#include <winsock2.h>
+#include <ws2tcpip.h>
+#else
+#include <netdb.h>
+#include <netinet/in.h>
+#include <netinet/tcp.h>
+#include <sys/select.h>
+#include <sys/socket.h>
+#include <sys/time.h>
+#include <unistd.h>
+#endif
+
+#if defined(_WIN32)
+#include <atomic>
+namespace {
+std::atomic<int> wsa_init_count = {0};
+} // anonymous namespace
+#else
+#include <fcntl.h>
+namespace {
+using SOCKET = int;
+} // anonymous namespace
+#endif
+
+namespace {
+constexpr SOCKET InvalidSocket = static_cast<SOCKET>(-1);
+void Init() {
+#if defined(_WIN32)
+ if (wsa_init_count++ == 0) {
+ WSADATA winsock_data;
+ (void)WSAStartup(MAKEWORD(2, 2), &winsock_data);
+ }
+#endif
+}
+
+void Term() {
+#if defined(_WIN32)
+ if (--wsa_init_count == 0) {
+ WSACleanup();
+ }
+#endif
+}
+
+bool SetBlocking(SOCKET s, bool blocking) {
+#if defined(_WIN32)
+ u_long mode = blocking ? 0 : 1;
+ return ioctlsocket(s, FIONBIO, &mode) == NO_ERROR;
+#else
+ auto arg = fcntl(s, F_GETFL, nullptr);
+ if (arg < 0) {
+ return false;
+ }
+ arg = blocking ? (arg & ~O_NONBLOCK) : (arg | O_NONBLOCK);
+ return fcntl(s, F_SETFL, arg) >= 0;
+#endif
+}
+
+bool Errored(SOCKET s) {
+ if (s == InvalidSocket) {
+ return true;
+ }
+ char error = 0;
+ socklen_t len = sizeof(error);
+ getsockopt(s, SOL_SOCKET, SO_ERROR, &error, &len);
+ return error != 0;
+}
+
+class Impl : public Socket {
+ public:
+ static std::shared_ptr<Impl> create(const char* address, const char* port) {
+ Init();
+
+ addrinfo hints = {};
+ hints.ai_family = AF_INET;
+ hints.ai_socktype = SOCK_STREAM;
+ hints.ai_protocol = IPPROTO_TCP;
+ hints.ai_flags = AI_PASSIVE;
+
+ addrinfo* info = nullptr;
+ auto err = getaddrinfo(address, port, &hints, &info);
+#if !defined(_WIN32)
+ if (err) {
+ printf("getaddrinfo(%s, %s) error: %s\n", address, port, gai_strerror(err));
+ }
+#else
+ (void)err;
+#endif
+
+ if (info) {
+ auto socket = ::socket(info->ai_family, info->ai_socktype, info->ai_protocol);
+ auto out = std::make_shared<Impl>(info, socket);
+ out->SetOptions();
+ return out;
+ }
+
+ Term();
+ return nullptr;
+ }
+
+ explicit Impl(SOCKET socket) : info(nullptr), s(socket) {}
+
+ Impl(addrinfo* i, SOCKET socket) : info(i), s(socket) {}
+
+ ~Impl() override {
+ if (info) {
+ freeaddrinfo(info);
+ }
+ Close();
+ Term();
+ }
+
+ template <typename FUNCTION>
+ void Lock(FUNCTION&& f) {
+ RLock l(mutex);
+ f(s, info);
+ }
+
+ void SetOptions() {
+ RLock l(mutex);
+ if (s == InvalidSocket) {
+ return;
+ }
+
+ int enable = 1;
+
+#if !defined(_WIN32)
+ // Prevent sockets lingering after process termination, causing
+ // reconnection issues on the same port.
+ setsockopt(s, SOL_SOCKET, SO_REUSEADDR, reinterpret_cast<char*>(&enable), sizeof(enable));
+
+ struct {
+ int l_onoff; /* linger active */
+ int l_linger; /* how many seconds to linger for */
+ } linger = {false, 0};
+ setsockopt(s, SOL_SOCKET, SO_LINGER, reinterpret_cast<char*>(&linger), sizeof(linger));
+#endif // !defined(_WIN32)
+
+ // Enable TCP_NODELAY.
+ setsockopt(s, IPPROTO_TCP, TCP_NODELAY, reinterpret_cast<char*>(&enable), sizeof(enable));
+ }
+
+ bool IsOpen() override {
+ {
+ RLock l(mutex);
+ if ((s != InvalidSocket) && !Errored(s)) {
+ return true;
+ }
+ }
+ WLock lock(mutex);
+ s = InvalidSocket;
+ return false;
+ }
+
+ void Close() override {
+ {
+ RLock l(mutex);
+ if (s != InvalidSocket) {
+#if defined(_WIN32)
+ closesocket(s);
+#else
+ ::shutdown(s, SHUT_RDWR);
+#endif
+ }
+ }
+
+ WLock l(mutex);
+ if (s != InvalidSocket) {
+#if !defined(_WIN32)
+ ::close(s);
+#endif
+ s = InvalidSocket;
+ }
+ }
+
+ size_t Read(void* buffer, size_t bytes) override {
+ {
+ RLock lock(mutex);
+ if (s == InvalidSocket) {
+ return 0;
+ }
+ auto len = recv(s, reinterpret_cast<char*>(buffer), bytes, 0);
+ if (len > 0) {
+ return static_cast<size_t>(len);
+ }
+ }
+ // Socket closed or errored
+ WLock lock(mutex);
+ s = InvalidSocket;
+ return 0;
+ }
+
+ bool Write(const void* buffer, size_t bytes) override {
+ RLock lock(mutex);
+ if (s == InvalidSocket) {
+ return false;
+ }
+ if (bytes == 0) {
+ return true;
+ }
+ return ::send(s, reinterpret_cast<const char*>(buffer), bytes, 0) > 0;
+ }
+
+ std::shared_ptr<Socket> Accept() override {
+ std::shared_ptr<Impl> out;
+ Lock([&](SOCKET socket, const addrinfo*) {
+ if (socket != InvalidSocket) {
+ Init();
+ if (auto sock = ::accept(socket, nullptr, nullptr); s >= 0) {
+ out = std::make_shared<Impl>(sock);
+ out->SetOptions();
+ }
+ }
+ });
+ return out;
+ }
+
+ private:
+ addrinfo* const info;
+ SOCKET s = InvalidSocket;
+ RWMutex mutex;
+};
+
+} // anonymous namespace
+
+Socket::~Socket() = default;
+
+std::shared_ptr<Socket> Socket::Listen(const char* address, const char* port) {
+ auto impl = Impl::create(address, port);
+ if (!impl) {
+ return nullptr;
+ }
+ impl->Lock([&](SOCKET socket, const addrinfo* info) {
+ if (bind(socket, info->ai_addr, info->ai_addrlen) != 0) {
+ impl.reset();
+ return;
+ }
+
+ if (listen(socket, 0) != 0) {
+ impl.reset();
+ return;
+ }
+ });
+ return impl;
+}
+
+std::shared_ptr<Socket> Socket::Connect(const char* address,
+ const char* port,
+ uint32_t timeout_ms) {
+ auto impl = Impl::create(address, port);
+ if (!impl) {
+ return nullptr;
+ }
+
+ std::shared_ptr<Socket> out;
+ impl->Lock([&](SOCKET socket, const addrinfo* info) {
+ if (socket == InvalidSocket) {
+ return;
+ }
+
+ if (timeout_ms == 0) {
+ if (::connect(socket, info->ai_addr, info->ai_addrlen) == 0) {
+ out = impl;
+ }
+ return;
+ }
+
+ if (!SetBlocking(socket, false)) {
+ return;
+ }
+
+ auto res = ::connect(socket, info->ai_addr, info->ai_addrlen);
+ if (res == 0) {
+ if (SetBlocking(socket, true)) {
+ out = impl;
+ }
+ } else {
+ const auto timeout_us = timeout_ms * 1000;
+
+ fd_set fdset;
+ FD_ZERO(&fdset);
+ FD_SET(socket, &fdset);
+
+ timeval tv;
+ tv.tv_sec = timeout_us / 1000000;
+ tv.tv_usec = static_cast<int>(timeout_us - (tv.tv_sec * 1000000));
+ res = select(static_cast<int>(socket + 1), nullptr, &fdset, nullptr, &tv);
+ if (res > 0 && !Errored(socket) && SetBlocking(socket, true)) {
+ out = impl;
+ }
+ }
+ });
+
+ if (!out) {
+ return nullptr;
+ }
+
+ return out->IsOpen() ? out : nullptr;
+}
diff --git a/src/tint/utils/socket/socket.h b/src/tint/utils/socket/socket.h
new file mode 100644
index 0000000..5ca37cc
--- /dev/null
+++ b/src/tint/utils/socket/socket.h
@@ -0,0 +1,71 @@
+// 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
+//
+// https://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_TINT_UTILS_SOCKET_SOCKET_H_
+#define SRC_TINT_UTILS_SOCKET_SOCKET_H_
+
+#include <atomic>
+#include <memory>
+
+/// Socket provides an OS abstraction to a TCP socket.
+class Socket {
+ public:
+ /// Connects to the given TCP address and port.
+ /// @param address the target socket address
+ /// @param port the target socket port
+ /// @param timeout_ms the timeout for the connection attempt.
+ /// If timeout_ms is non-zero and no connection was made before timeout_ms milliseconds,
+ /// then nullptr is returned.
+ /// @returns the connected Socket, or nullptr on failure
+ static std::shared_ptr<Socket> Connect(const char* address,
+ const char* port,
+ uint32_t timeout_ms);
+
+ /// Begins listening for connections on the given TCP address and port.
+ /// Call Accept() on the returned Socket to block and wait for a connection.
+ /// @param address the socket address to listen on. Use "localhost" for connections from only
+ /// this machine, or an empty string to allow connections from any incoming address.
+ /// @param port the socket port to listen on
+ /// @returns the Socket that listens for connections
+ static std::shared_ptr<Socket> Listen(const char* address, const char* port);
+
+ /// Destructor
+ virtual ~Socket();
+
+ /// Attempts to read at most `n` bytes into buffer, returning the actual number of bytes read.
+ /// read() will block until the socket is closed or at least one byte is read.
+ /// @param buffer the output buffer. Must be at least `n` bytes in size.
+ /// @param n the maximum number of bytes to read
+ /// @return the number of bytes read, or 0 if the socket was closed or errored
+ virtual size_t Read(void* buffer, size_t n) = 0;
+
+ /// Writes `n` bytes from buffer into the socket.
+ /// @param buffer the source data buffer. Must be at least `n` bytes in size.
+ /// @param n the number of bytes to read from `buffer`
+ /// @returns true on success, or false if there was an error or the socket was
+ /// closed.
+ virtual bool Write(const void* buffer, size_t n) = 0;
+
+ /// @returns true if the socket has not been closed.
+ virtual bool IsOpen() = 0;
+
+ /// Closes the socket.
+ virtual void Close() = 0;
+
+ /// Blocks for a connection to be made to the listening port, or for the Socket to be closed.
+ /// @returns a pointer to the next established incoming connection
+ virtual std::shared_ptr<Socket> Accept() = 0;
+};
+
+#endif // SRC_TINT_UTILS_SOCKET_SOCKET_H_
