diff --git a/.github/workflows/ci.yml b/.github/workflows/ci.yml index cca1bd3..9454261 100644 --- a/.github/workflows/ci.yml +++ b/.github/workflows/ci.yml @@ -18,12 +18,18 @@ env: # saved and restored `~/.mcpp` -- a directory mcpp never touched. Every run # was therefore the cold case, which is what the 90-minute budget below was # sized for. mcpp's own ci-linux.yml pins it for this reason and says so. - MCPP_HOME: /home/runner/.mcpp + # + # Set per JOB rather than here: the path differs by platform, and a + # workflow-level `/home/runner/.mcpp` would silently be wrong on the other + # two runners -- wrong in the direction that still appears to work, which is + # the kind this repository has already paid for once. jobs: consumers: name: consumers (linux x86_64) runs-on: ubuntu-24.04 + env: + MCPP_HOME: /home/runner/.mcpp # 90 rather than 60: the SYCL fixture pulls three payloads this job did not # need before -- dpcpp (578 MB installed), gcc (265 MB) and cuda-nvcc # (319 MB) -- and a cold cache downloads all of them before the first @@ -90,6 +96,49 @@ jobs: grep -q '^magic=07230203' run-glslc.log rm -rf shaderc.tar.gz shaderc-2026.3 + # TWO SHADERS THAT DIFFER ONLY BY DIRECTORY. + # + # The output name is the stem and the stage, so `a/x.vert` and `b/x.vert` + # both produce `x_vert.h` declaring `x_vert_spv`. Before this rule checked + # for it, ninja caught the duplicate output -- so it was never silent, but + # the message named the generated file and neither shader, arrived as a + # graph-loading failure rather than as this rule's refusal, and stated no + # way out. A project with one shader per stage never meets it, which is + # how it survived; a graphics project organising shaders by purpose is the + # first to have two. + - name: two shaders with one stem are refused, naming both + working-directory: tests/spirv-consumer + run: | + set -e + mkdir -p shaders/dup + cp shaders/scale.comp shaders/dup/scale.comp + # The fixture globs `shaders/*.comp`; widen it for this step only. + cp mcpp.toml mcpp.toml.bak + sed -i 's#shaders/\*.comp#shaders/**/*.comp#' mcpp.toml + rm -rf target + if "$MCPP" build > dup.log 2>&1; then + echo "FAIL: two shaders mapping to one output were accepted" + exit 1 + fi + for needle in 'two shaders map to one output' 'shaders/scale.comp' \ + 'shaders/dup/scale.comp' 'scale_comp_spv' 'rename one of them'; do + grep -q -- "$needle" dup.log || { + echo "FAIL: the refusal does not mention '$needle'" + tail -20 dup.log + exit 1 + } + done + echo "ok: refused, naming both shaders, the symbol and the way out" + # …and the fixture still builds once the duplicate is gone, so this + # step cannot leave a project that refuses everything. + mv mcpp.toml.bak mcpp.toml + rm -rf shaders/dup target + "$MCPP" build > restored.log 2>&1 || { + echo "FAIL: the fixture no longer builds after the duplicate was removed" + tail -20 restored.log; exit 1; } + rm -f dup.log restored.log + echo "ok: and it builds again with one shader per stem" + # A tool, not a rule: the header is written while the build program runs, # so there is no action to schedule. The second build is the measurement # that matters -- editing the data file must reach the binary, which is @@ -144,6 +193,21 @@ jobs: "$MCPP" run --no-accel | tee run.log grep -q '^12 24 36 48' run.log + # EVERY RULE COMPILED FOR THIS HOST, which is a different claim from any + # step above. Each of those exercises one rule and needs that rule's + # payload, so a rule is only ever compiled on the platforms its payload + # exists for -- and a rule's host-dependent halves are exactly the code + # those platforms do not compile. This fixture names no accelerator, so + # every rule returns before looking for a payload and nothing is + # downloaded; what it asserts is that all six modules compile here. + - name: every rule module compiles for this host + working-directory: tests/all-rules-compile + run: | + set -e + "$MCPP" build + "$MCPP" run | tee run.log + grep -q '^all-rules-compile ok' run.log + # TWO RULES IN ONE BUILD PROGRAM, which every step above has exactly # one of. With one, a rule can take the whole of `device_sources()` and # be right by accident; with two, that list holds a `.cu` AND a `.comp`, @@ -193,8 +257,14 @@ jobs: # rehearsal passed because a DEV build is disqualified from # self-contained mode and does use ~/.mcpp, so the rehearsal and CI # were clearing different directories. - HOME_DIR=$("$MCPP" self env | awk -F'= *' '/^MCPP_HOME/{print $2; exit}') - [ -n "$HOME_DIR" ] || { echo "FAIL: could not read MCPP_HOME"; exit 1; } + # Through a FILE rather than a pipe: on macOS this same expression + # killed mcpp with `internal: unhandled exception: failed to write + # formatted output`, because `awk ... exit` closes the read end at + # the first match. Written the same way on every platform so the two + # jobs cannot drift. + "$MCPP" self env > mcpp-env.txt + HOME_DIR=$(awk -F'= *' '/^MCPP_HOME/{print $2; exit}' mcpp-env.txt) + [ -n "$HOME_DIR" ] || { echo "FAIL: could not read MCPP_HOME"; cat mcpp-env.txt; exit 1; } echo "clearing provisioning stamps under $HOME_DIR" rm -rf "$HOME_DIR/provisioned" for d in tests/cuda-consumer tests/hip-consumer \ @@ -281,3 +351,162 @@ jobs: fi done [ "$fail" -eq 0 ] + + # ── THE SAME RULE ON THE OTHER TWO PLATFORMS ──────────────────────────────── + # + # `rules-spirv` is the only member whose payload this ecosystem publishes for + # all three, so it is the only one that can be checked on all three. What this + # job proves is the claim the per-platform declaration makes: a project writes + # the rule edge and nothing else, and the compiler that arrives is whichever + # one that platform has -- glslang on Linux, glslc here. + # + # ONLY THE SPIR-V FIXTURE. The CUDA, HIP and SYCL payloads are Linux-only + # upstream, so a matrix that ran every fixture everywhere would be red for a + # reason that is not a defect. + rules-cross-platform: + name: rules (${{ matrix.name }}) + runs-on: ${{ matrix.runs-on }} + timeout-minutes: 60 + strategy: + fail-fast: false + matrix: + include: + # macos-15, matching mcpp's own macOS CI, and the reason is a + # measured limitation of the older release rather than a preference. + # + # On macos-14 a build program fails to LINK, with + # + # ld64.lld: error: undefined symbol: + # std::__1::__is_posix_terminal(__sFILE*) + # + # referenced from `std::__1::__print::__is_terminal`. That symbol + # lives in the libc++ DYLIB and was added in a version macOS 14 does + # not ship, and the build program's link resolves `-lc++` to the + # system copy rather than to the LLVM payload's. It is reached + # through `import std` itself, so no rule package can avoid it -- the + # rules stopped using `std::println` for the same reason and it was + # not enough. + # + # Recorded rather than worked around here: the engine-side half is + # `host_link_tokens` returning before it adds the payload's runtime + # directories on macOS. Once a release carries that, this moves back + # to macos-14, which is the floor mcpp's README states. + - name: macos arm64 + runs-on: macos-15 + asset: macosx-arm64.tar.gz + dir-suffix: macosx-arm64 + - name: windows x86_64 + runs-on: windows-2022 + asset: windows-x86_64.zip + dir-suffix: windows-x86_64 + defaults: + run: + shell: bash + steps: + - uses: actions/checkout@v4 + + - name: Pin MCPP_HOME under this runner's home + run: | + set -e + # The value has to be in the HOST's path syntax, not Git Bash's. + # `$HOME` there is `/c/Users/runneradmin`, and mcpp hands this + # variable to xlings, which is a native Windows program: it answered + # `The filename, directory name, or volume label syntax is + # incorrect.` and exited 1, from inside a provisioning step whose own + # message named a package rather than a path. + H="$HOME/.mcpp" + command -v cygpath >/dev/null && H=$(cygpath -m "$H") + echo "MCPP_HOME=$H" >> "$GITHUB_ENV" + + - name: Cache the mcpp sandbox + uses: actions/cache@v4 + with: + path: ~/.mcpp + key: mcpp-sandbox-${{ runner.os }}-${{ runner.arch }}-${{ env.MCPP_VERSION }}-${{ hashFiles('tests/spirv-consumer/mcpp.toml') }} + restore-keys: | + mcpp-sandbox-${{ runner.os }}-${{ runner.arch }}-${{ env.MCPP_VERSION }}- + + - name: Fetch the released mcpp + run: | + set -e + curl -L -fsS --retry 3 --retry-all-errors -o mcpp.pkg \ + "https://github.com/mcpp-community/mcpp/releases/download/v${MCPP_VERSION}/mcpp-${MCPP_VERSION}-${{ matrix.asset }}" + case "${{ matrix.asset }}" in + *.zip) unzip -q mcpp.pkg ;; + *) tar -xzf mcpp.pkg ;; + esac + dir="mcpp-${MCPP_VERSION}-${{ matrix.dir-suffix }}" + MCPP="$PWD/$dir/bin/mcpp" + "$MCPP" --version + # The bundled xlings, named explicitly because MCPP_HOME is pinned + # away from the tarball. See the Linux job for the measurement. + # + # The suffix is the host's: the Windows distribution ships + # `registry/bin/xlings.exe`, and a variable pointing at a file that + # does not exist produces `error: xlings binary not found` naming + # three remedies, none of which is the one that applies. + # Two host differences, and the second one is the interesting one. + # The suffix: the Windows distribution ships `xlings.exe`. The PATH + # SHAPE: this job runs under Git Bash, whose `$PWD` is a POSIX path + # (`/d/a/...`), while `mcpp` is a native Windows program and reads + # this variable as a Windows path. Both failures report the same + # `error: xlings binary not found` and offer the same three remedies, + # none of which is "the path you gave is in the other syntax". + XL="$PWD/$dir/registry/bin/xlings" + [ -f "$XL" ] || XL="$XL.exe" + test -f "$XL" || { echo "no vendored xlings at $XL"; exit 1; } + command -v cygpath >/dev/null && XL=$(cygpath -m "$XL") + export MCPP_VENDORED_XLINGS="$XL" + "$MCPP" self config --mirror GLOBAL + echo "MCPP=$MCPP" >> "$GITHUB_ENV" + echo "MCPP_VENDORED_XLINGS=$MCPP_VENDORED_XLINGS" >> "$GITHUB_ENV" + + # FIRST, BECAUSE IT IS THE CHEAPER QUESTION AND THE MORE INFORMATIVE + # ANSWER. Every other step here drives one rule end to end and needs + # that rule's payload; this one names no accelerator, downloads nothing, + # and asks only whether all six modules COMPILE for this host -- which + # is the half of a rule that a Linux-only CI never sees. Run after the + # end-to-end step, its failure arrived as somebody else's build error: + # `no member named 'popen' in the global namespace`, reported against a + # consumer that had nothing to do with it. + - name: every rule module compiles for this host + working-directory: tests/all-rules-compile + run: | + set -e + "$MCPP" build + "$MCPP" run | tee run.log + grep -q '^all-rules-compile ok' run.log + + - name: rules-spirv through a consumer + working-directory: tests/spirv-consumer + run: | + set -e + "$MCPP" build + "$MCPP" run | tee run.log + grep -q '^magic=07230203' run.log + + # The same provenance criterion the Linux job applies: the payload has to + # come from the GRAPH, not from the fixture. Without it, a fixture that + # quietly regained an `[xlings.workspace]` would keep this green while the + # claim stopped being true. + - name: the rule declared its own compiler + working-directory: tests/spirv-consumer + run: | + set -e + if grep -qE '^\[(target\..*\.)?xlings\.workspace\]' mcpp.toml; then + echo "FAIL: the fixture declares payloads itself"; exit 1 + fi + # `mcpp self env > file`, not `mcpp self env | awk`. Measured on + # macos-15: through a pipe whose reader exits at the first match, + # mcpp died with `internal: unhandled exception: failed to write + # formatted output` and exit 70. A file has no reader to disappear. + "$MCPP" self env > env.txt + HOME_DIR=$(awk -F'= *' '/^MCPP_HOME/{print $2; exit}' env.txt) + test -n "$HOME_DIR" || { echo "FAIL: mcpp self env named no MCPP_HOME"; cat env.txt; exit 1; } + rm -rf "$HOME_DIR/provisioned" target + "$MCPP" build > prov.log 2>&1 + grep -q 'entries declared by dependencies' prov.log || { + echo "FAIL: no payload came from the graph" + grep -i provisioning prov.log || echo "(no provisioning line at all)" + exit 1; } + echo "ok: $(grep -m1 'entries declared by dependencies' prov.log)" diff --git a/mcpp.toml b/mcpp.toml index 5ad0d38..d53832c 100644 --- a/mcpp.toml +++ b/mcpp.toml @@ -1,7 +1,7 @@ [package] name = "plugins" namespace = "mcpp" -version = "0.2.4" +version = "0.2.5" description = "Official mcpp build plugins: rule packages under mcpp.rules.*, build-time utilities under mcpp.tools.*, each member selected by a feature" license = "Apache-2.0" authors = ["mcpp-community"] @@ -112,15 +112,62 @@ tools-embed = { sources = ["tools/embed.cppm"] } # SET, so this is a conjunction of two membership tests rather than a # contradiction -- and without it every SYCL build would download a CUDA # toolkit it may have no target for. -[target.'cfg(all(accelerator = "sycl", accelerator = "cuda"))'.feature-xlings.rules-sycl] +# +# `linux` is part of the predicate for a reason that is measured rather than +# stylistic: upstream does not build the CUDA plugin for Windows, and the +# published Windows asset carries Level Zero and OpenCL adapters only. The rule +# refuses that combination on that host, and a refusal that arrives after a +# multi-gigabyte download is a worse refusal -- provisioning runs before the +# rule does. +[target.'cfg(all(accelerator = "sycl", accelerator = "cuda", linux))'.feature-xlings.rules-sycl] "xim:cuda-nvcc" = "12.9.86" -# glslang alone: it is the rule's first choice, and `xim:shaderc` is the -# fallback a project names when it wants glslc instead. Declaring both would -# install both and use one. -[target.'cfg(accelerator = "vulkan")'.feature-xlings.rules-spirv] +# ONE COMPILER PER PLATFORM, AND NOT THE SAME ONE EVERYWHERE. +# +# This rule drives both reference compilers and writes the C declaration itself +# on the glslc route, so which one is present is not a difference a consumer +# sees. That is what makes cross-platform parity affordable: it is provided by +# the rule's ability to CHOOSE, not by publishing one compiler three times. +# +# Linux keeps `xim:glslang`, so nothing about an existing Linux build changes. +# macOS and Windows take `xim:shaderc`, which is the compiler this ecosystem +# publishes for them -- glslang has no upstream binary for either, and building +# it three times buys nothing glslc does not already give. +# +# Declaring both on one platform would install both and use one, so each block +# names exactly the compiler that platform will run. +[target.'cfg(all(accelerator = "vulkan", linux))'.feature-xlings.rules-spirv] "xim:glslang" = ">=15.1.0" +# AN EXACT VERSION HERE, WHERE LINUX HAS A FLOOR, AND THE REASON IS THE +# ENGINE'S COMMAND LINE RATHER THAN THE PACKAGE. +# +# mcpp hands its provisioning request to xlings as a JSON argument on a shell +# command line. On Windows that line is parsed by cmd.exe, whose quoting rules +# are not the ones the JSON is escaped for: the `\"` sequences leave cmd's +# quote state OFF by the time it reaches a `>` in a version constraint, so the +# `>` is read as a REDIRECTION. Measured on windows-2022 with mcpp 2026.9.6.6: +# +# Provisioning [xlings.workspace] entries declared by dependencies +# (xim:shaderc@>=2026.3) +# The filename, directory name, or volume label syntax is incorrect. +# error: ... xlings exited 1 +# +# -- cmd's own message about an unusable redirection target, arriving as a +# packaging failure. No declaration reachable on Windows had carried a `>` +# before, which is why the whole `>=` shape had never been exercised there. +# +# An exact version is a legitimate declaration and not a workaround in +# disguise: mcpp reads it as a CHOICE, so a project that wants a different one +# still wins and the override is reported. macOS takes the same value so the +# two platforms that use this compiler agree. Both revert to `>=2026.3` once a +# released engine escapes the argument for cmd. +[target.'cfg(all(accelerator = "vulkan", macos))'.feature-xlings.rules-spirv] +"xim:shaderc" = "2026.3" + +[target.'cfg(all(accelerator = "vulkan", windows))'.feature-xlings.rules-spirv] +"xim:shaderc" = "2026.3" + # 8.5.0 is a real floor rather than a preference: the mixed-mode object -- the # one carrying both the device binary and a host-callable launcher, which is # what lets it join an ordinary link -- and the 38 SoC simulators are what this diff --git a/rules/ascendc.cppm b/rules/ascendc.cppm index 43bef15..b7325d4 100644 --- a/rules/ascendc.cppm +++ b/rules/ascendc.cppm @@ -59,6 +59,25 @@ export module mcpp.rules.ascendc; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::rules::ascendc { // ─── What the engine said ────────────────────────────────────────────────── @@ -209,8 +228,7 @@ inline std::optional find_toolkit() { toolkit t; const auto pkg = xpkg("cann-toolkit"); if (pkg.empty()) { - std::println(std::cerr, - "mcpp.rules.ascendc: the CANN toolkit is not installed.\n" + std::cerr << std::format("mcpp.rules.ascendc: the CANN toolkit is not installed.\n" " This rule DECLARES it, so a project normally writes nothing. Check, in " "order:\n" " mcpp older than 2026.9.6.6; `features = [\"rules-ascendc\"]` missing from " @@ -220,7 +238,7 @@ inline std::optional find_toolkit() { " [target.'cfg(accelerator = \"ascend\")'.xlings.workspace]\n" " \"xim:cann-toolkit\" = \"8.5.0\"\n" " It carries both halves this rule needs: the device compiler and,\n" - " for a machine with no NPU, the per-SoC simulators."); + " for a machine with no NPU, the per-SoC simulators.") << '\n'; return std::nullopt; } t.root = pkg + "/cann"; @@ -229,10 +247,9 @@ inline std::optional find_toolkit() { if (std::filesystem::is_directory(candidate)) { t.arch_root = candidate; break; } } if (t.arch_root.empty() || !std::filesystem::exists(t.bisheng())) { - std::println(std::cerr, - "mcpp.rules.ascendc: `xim:cann-toolkit` is installed at '{}' but has no\n" + std::cerr << std::format("mcpp.rules.ascendc: `xim:cann-toolkit` is installed at '{}' but has no\n" " device compiler under -linux/ccec_compiler/bin/bisheng.\n" - " The install is incomplete; reinstall the package.", pkg); + " The install is incomplete; reinstall the package.", pkg) << '\n'; return std::nullopt; } return t; @@ -258,8 +275,7 @@ inline std::vector plan(std::span sources, options opt std::vector out; const std::string root = mcpp::manifest_dir(); if (root.empty()) { - std::println(std::cerr, - "mcpp.rules.ascendc: no mcpp build context -- this runs from build.mcpp"); + std::cerr << std::format("mcpp.rules.ascendc: no mcpp build context -- this runs from build.mcpp") << '\n'; return out; } const auto tg = parse_target(mcpp::accel()); @@ -267,12 +283,11 @@ inline std::vector plan(std::span sources, options opt // The same refusal `mcpp.rules.cuda` makes, for the same reason: a // device build that names no device is refused HERE rather than at run // time, where it is a kernel that does not exist for the part present. - std::println(std::cerr, - "mcpp.rules.ascendc: [build] accel names no Da Vinci architecture " + std::cerr << std::format("mcpp.rules.ascendc: [build] accel names no Da Vinci architecture " "(accel = \"{}\").\n" " Write e.g. accel = \"ascend8.5+{{dav-c220}}\" -- the set a build\n" " compiles for is a decision, and the machine's own hardware is a poor\n" - " default for it.", mcpp::accel()); + " default for it.", mcpp::accel()) << '\n'; return out; } auto tk = find_toolkit(); diff --git a/rules/cuda.cppm b/rules/cuda.cppm index 4fc7318..27c4c44 100644 --- a/rules/cuda.cppm +++ b/rules/cuda.cppm @@ -45,8 +45,44 @@ export module mcpp.rules.cuda; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::rules::cuda { +// A PROGRAM'S NAME CARRIES ITS HOST'S SUFFIX, AND THE PAYLOAD LAYOUT DOES TOO. +// +// The NVIDIA redistributables are published for Windows as well as Linux and +// carry the same tree under different names: `bin/nvcc.exe` rather than +// `bin/nvcc`, and import libraries under `lib/x64` rather than shared objects +// under `lib` or `lib64`. Every one of the paths this rule builds is derived +// from those two facts, so both are stated once here rather than at each site +// -- a `std::filesystem::exists` against the wrong spelling does not fail, it +// answers false, and this rule reads several such answers as decisions. +#if defined(_WIN32) +inline constexpr bool kWindows = true; +inline constexpr const char* kExe = ".exe"; +#else +inline constexpr bool kWindows = false; +inline constexpr const char* kExe = ""; +#endif + enum class route { automatic, clang, nvcc }; struct options { @@ -177,7 +213,7 @@ struct toolkit { // 12.x toolkits ship as the separate `cuda-cccl` package, and which the // host's /usr/include had supplied in the same way. std::string curand_root, cccl_root; - std::string nvcc() const { return nvcc_root + "/bin/nvcc"; } + std::string nvcc() const { return nvcc_root + "/bin/nvcc" + kExe; } std::string host_config() const { for (auto const* r : { &crt_root, &nvcc_root, &cudart_root }) { if (r->empty()) continue; @@ -203,8 +239,11 @@ struct toolkit { } std::vector lib_dirs() const { std::vector out; + // `/lib/x64` is the Windows layout; it is searched on every host + // because a directory that does not exist contributes nothing, and a + // list keyed on the host would be a second place to keep in step. for (auto const* r : { &cudart_root, &nvcc_root }) - for (auto const* sub : { "/lib", "/lib64" }) + for (auto const* sub : { "/lib", "/lib64", "/lib/x64" }) if (!r->empty() && std::filesystem::is_directory(*r + sub)) out.push_back(*r + sub); return out; @@ -225,8 +264,7 @@ inline std::optional find_toolkit() { t.cccl_root = xpkg("cuda-cccl"); t.driver_dir = xpkg("libcuda-host-link"); if (t.nvcc_root.empty() || t.cudart_root.empty()) { - std::println(std::cerr, - "mcpp.rules.cuda: the toolkit is not installed.\n" + std::cerr << std::format("mcpp.rules.cuda: the toolkit is not installed.\n" " This rule DECLARES it, so a project normally writes nothing. Three " "things stop\n" " that from reaching the build, in the order worth checking:\n" @@ -240,7 +278,7 @@ inline std::optional find_toolkit() { " [target.'cfg(accelerator = \"cuda\")'.xlings.workspace]\n" " \"xim:cuda-nvcc\" = \"12.9.86\"\n" " \"xim:cuda-cudart\" = \"12.9.79\"\n" - " (found nvcc: '{}', cudart: '{}')", t.nvcc_root, t.cudart_root); + " (found nvcc: '{}', cudart: '{}')", t.nvcc_root, t.cudart_root) << '\n'; return std::nullopt; } return t; @@ -451,27 +489,42 @@ struct edge { std::vector command, inputs, outputs; }; +// ON WINDOWS THE DEFAULT IS THE CLANG ROUTE WHATEVER THE PROJECT'S COMPILER IS. +// +// The two routes differ in who compiles the host half of a device unit. On the +// nvcc route that is a compiler nvcc drives through `-ccbin`, and on Windows +// the only one it accepts is MSVC's `cl.exe` -- whose directory is found +// through `vswhere` and the installed Visual Studio instance, neither of which +// mcpp resolves or states to a build program. The rule would have to search +// the host for it, which is the one thing a rule of this ecosystem does not do. +// +// The clang route has no such gap: it drives no second compiler, and clang +// locates the MSVC headers and libraries it needs for the host half itself, +// the same way it does for every ordinary C++ translation unit on this host. inline route decide(route asked) { if (asked != route::automatic) return asked; +#if defined(_WIN32) + return route::clang; +#else return std::string_view(mcpp::compiler()) == "clang" ? route::clang : route::nvcc; +#endif } inline std::vector plan(std::span sources, options opt = {}) { std::vector out; const std::string root = mcpp::manifest_dir(); if (root.empty()) { - std::println(std::cerr, "mcpp.rules.cuda: no mcpp build context -- this runs from build.mcpp"); + std::cerr << std::format("mcpp.rules.cuda: no mcpp build context -- this runs from build.mcpp") << '\n'; return out; } const auto tg = parse_target(mcpp::accel()); if (!tg.present || tg.archs.empty()) { // C19: a device build that names no device is refused HERE, not at // run time as `no kernel image is available for execution`. - std::println(std::cerr, - "mcpp.rules.cuda: [build] accel names no CUDA architecture (accel = \"{}\").\n" + std::cerr << std::format("mcpp.rules.cuda: [build] accel names no CUDA architecture (accel = \"{}\").\n" " Write e.g. accel = \"cuda12.9+{{sm_89}} ptx>=89\" -- the set a build compiles\n" " for is a decision, and the machine's own hardware is a poor default for it.", - mcpp::accel()); + mcpp::accel()) << '\n'; return out; } auto tk = find_toolkit(); @@ -483,9 +536,17 @@ inline std::vector plan(std::span sources, options opt std::string driver_cc; // the compiler that runs the device unit std::vector front; // the command up to the input file if (r == route::clang) { - driver_cc = tcdir + "/bin/clang++"; + driver_cc = tcdir + "/bin/clang++" + kExe; if (!std::filesystem::exists(driver_cc)) { - std::println(std::cerr, "mcpp.rules.cuda: the clang route needs the toolchain's clang++ at {}", driver_cc); + std::cerr << std::format("mcpp.rules.cuda: the clang route needs the toolchain's clang++ at {}.\n" + " Name an LLVM toolchain for this project:\n" + " [toolchain]\n" + " default = \"llvm@22.1.8\"{}", + driver_cc, + kWindows ? "\n On Windows this route is the only one the rule takes: nvcc " + "drives a host\n compiler named by -ccbin, and locating MSVC's " + "cl.exe is not something\n this rule does." + : "") << '\n'; return out; } // Refused here rather than at clang's include error: the header it @@ -495,18 +556,17 @@ inline std::vector plan(std::span sources, options opt const bool curand_ok = !tk->curand_root.empty() && std::filesystem::exists(tk->curand_root + "/include/curand_mtgp32_kernel.h"); if (!curand_ok || !tk->has_cccl()) { - std::println(std::cerr, - "mcpp.rules.cuda: the clang route needs cuRAND's headers, which clang's CUDA " + std::cerr << std::format("mcpp.rules.cuda: the clang route needs cuRAND's headers, which clang's CUDA " "wrapper includes unconditionally, and CCCL's, which they include in turn.\n" " This rule declares both; see the note above for why they may not have " "arrived.\n" " To pin a different line, name it in your own project and it wins:\n" " \"xim:cuda-cccl\" = \"12.9.27\" (the 12.9 line; 13.x pairs with 13.x)\n" " \"xim:libcurand\" = \"10.3.10.19\" (the 12.9 line; 10.4.x pairs with 13.x)\n" - " (found cccl: '{}', curand: '{}')", tk->cccl_root, tk->curand_root); + " (found cccl: '{}', curand: '{}')", tk->cccl_root, tk->curand_root) << '\n'; return out; } - front = { driver_cc, "-x", "cuda", "-std=c++17", "-O2", "-fPIC", + front = { driver_cc, "-x", "cuda", "-std=c++17", "-O2", "--cuda-path=" + tk->nvcc_root, "-Wno-unknown-cuda-version", // NVIDIA'S HEADER REFUSES libc++, AND THE REFUSAL IS // ABOUT nvcc RATHER THAN ABOUT THIS COMPILER. @@ -527,15 +587,35 @@ inline std::vector plan(std::span sources, options opt // on this route: nvcc's host pass really does break against // libc++, and nothing here weakens that. "-D_ALLOW_UNSUPPORTED_LIBCPP" }; + // Position-independent code is the default on Windows and naming it + // is an unused-argument warning on every device unit. + if constexpr (!kWindows) front.insert(front.begin() + 5, "-fPIC"); for (auto const& inc : tk->include_dirs()) front.push_back("-I" + inc); for (auto const& a : tg.archs) front.push_back("--cuda-gpu-arch=" + a); // clang checks ptxas and fatbinary itself; say so before it does. for (auto const* tool : { "ptxas", "fatbinary" }) - if (!std::filesystem::exists(tk->nvcc_root + "/bin/" + tool)) + if (!std::filesystem::exists(tk->nvcc_root + "/bin/" + tool + kExe)) mcpp::warning(std::format("the toolkit payload has no {}; clang invokes it " "after generating PTX", tool).c_str()); - std::println("mcpp.rules.cuda: clang route -- {} (toolkit {})", driver_cc, tk->nvcc_root); + std::cout << std::format("mcpp.rules.cuda: clang route -- {} (toolkit {})", driver_cc, tk->nvcc_root) << '\n'; } else { + // ASKED FOR EXPLICITLY, BECAUSE `decide` NEVER CHOOSES IT HERE. + // Everything below reads a GCC bound out of the toolkit's header and + // hands nvcc a `g++` to drive. On Windows there is no such compiler in + // the toolchain and the bound in `host_config.h` is stated in + // `_MSC_VER`, so the branch would build a command line out of two + // answers that mean nothing and nvcc would report the third. + if constexpr (kWindows) { + std::cerr << std::format("mcpp.rules.cuda: the nvcc route was asked for, and on Windows this rule does " + "not take it.\n" + " nvcc compiles the host half through a compiler named by -ccbin, which on " + "this host\n is MSVC's cl.exe; finding it means asking the machine about its " + "Visual Studio\n installation, which is the kind of host dependence this " + "ecosystem removes.\n" + " Leave the route unset: the clang route is the default here and needs no " + "second compiler.") << '\n'; + return out; + } // nvcc drives the toolchain's own compiler, and refuses one newer than // the bound its header states. Read the bound; if exceeded, pass the // escape hatch and say so -- an unexplained flag is worse than a note. @@ -546,10 +626,9 @@ inline std::vector plan(std::span sources, options opt // LLVM toolchain's clang uses. The pairing that works is nvcc with // a GCC toolchain; with an LLVM toolchain the clang route is the // one to take, and it is the default. - std::println(std::cerr, - "mcpp.rules.cuda: the nvcc route needs a GCC host compiler; this project's " + std::cerr << std::format("mcpp.rules.cuda: the nvcc route needs a GCC host compiler; this project's " "toolchain is LLVM, whose clang uses libc++ and nvcc refuses it. Use the clang " - "route (the default for an LLVM toolchain) or set [toolchain] to a gcc payload."); + "route (the default for an LLVM toolchain) or set [toolchain] to a gcc payload.") << '\n'; return out; } // The other pairing this route cannot have: an old toolkit and a C @@ -558,8 +637,7 @@ inline std::vector plan(std::span sources, options opt // decision. if (major_of(tg.version) < 13 && libc_declares_c23_pi_math(mcpp::toolchain_sysroot())) { - std::println(std::cerr, - "mcpp.rules.cuda: toolkit {} redeclares the C23 functions cospi, sinpi and " + std::cerr << std::format("mcpp.rules.cuda: toolkit {} redeclares the C23 functions cospi, sinpi and " "rsqrt for the host without `noexcept`, and the C library this build compiles " "against declares them with it; nvcc's front end refuses the pair.\n" " Name a 13.x toolkit, whose headers leave them to the C library:\n" @@ -568,7 +646,7 @@ inline std::vector plan(std::span sources, options opt " \"xim:cuda-crt\" = \"13.3.33\"\n" " \"xim:cuda-cudart\" = \"13.3.29\"\n" " or take the clang route, which does not include that header at all.", - tg.version); + tg.version) << '\n'; return out; } // The host compiler nvcc drives, chosen within the bound the toolkit @@ -580,27 +658,26 @@ inline std::vector plan(std::span sources, options opt // otherwise a gcc payload the project declared for this purpose, and // otherwise a refusal that says which declaration to add. const auto b = read_bounds(tk->host_config()); - const std::string tcGcc = tcdir + "/bin/g++"; + const std::string tcGcc = tcdir + "/bin/g++" + kExe; const int tcMajor = compiler_major(tcGcc); if (b.gcc == 0 || tcMajor <= b.gcc) { driver_cc = tcGcc; } else if (auto payload = xpkg("gcc"); !payload.empty() - && compiler_major(payload + "/bin/g++") <= b.gcc) { - driver_cc = payload + "/bin/g++"; + && compiler_major(payload + "/bin/g++" + kExe) <= b.gcc) { + driver_cc = payload + "/bin/g++" + kExe; mcpp::warning(std::format( "nvcc {} states gcc <= {} in {}; the toolchain's gcc {} exceeds it, so the " "device unit is compiled with the declared xim:gcc payload ({}). The clang " "route has no such bound.", tg.version, b.gcc, tk->host_config(), tcMajor, driver_cc).c_str()); } else { - std::println(std::cerr, - "mcpp.rules.cuda: nvcc {} accepts gcc <= {} ({}), and this project's " + std::cerr << std::format("mcpp.rules.cuda: nvcc {} accepts gcc <= {} ({}), and this project's " "toolchain is gcc {}.\n" " Declare a gcc payload within the bound and the rule drives that one:\n" " [xlings.workspace]\n" " \"xim:gcc\" = \"13.3.0\"\n" " or take the clang route with [toolchain] default = \"llvm@22.1.8\".", - tg.version, b.gcc, tk->host_config(), tcMajor); + tg.version, b.gcc, tk->host_config(), tcMajor) << '\n'; return out; } front = { tk->nvcc(), "-ccbin", driver_cc, "-std=c++17", "-O2", @@ -634,7 +711,7 @@ inline std::vector plan(std::span sources, options opt "nvcc cannot reach its own back-end: it invokes '{}' by name and that name " "does not resolve on the search path it states. On the 13.x line install " "xim:libnvvm beside xim:cuda-nvcc.", *missing).c_str()); - std::println("mcpp.rules.cuda: nvcc route -- {} with -ccbin {}", tk->nvcc(), driver_cc); + std::cout << std::format("mcpp.rules.cuda: nvcc route -- {} with -ccbin {}", tk->nvcc(), driver_cc) << '\n'; } // The link line gets its directories from here, not from the manifest: the diff --git a/rules/hip.cppm b/rules/hip.cppm index f4388ac..eb5a8dd 100644 --- a/rules/hip.cppm +++ b/rules/hip.cppm @@ -38,6 +38,25 @@ export module mcpp.rules.hip; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::rules::hip { // The two implementations, named. `automatic` reads the accelerator axis. @@ -222,8 +241,7 @@ inline std::vector plan(std::span sources, options opt std::vector out; const std::string root = mcpp::manifest_dir(); if (root.empty()) { - std::println(std::cerr, - "mcpp.rules.hip: no mcpp build context -- this runs from build.mcpp"); + std::cerr << std::format("mcpp.rules.hip: no mcpp build context -- this runs from build.mcpp") << '\n'; return out; } @@ -238,27 +256,25 @@ inline std::vector plan(std::span sources, options opt // runtime and device library, and this ecosystem publishes neither // yet; compiling for it would produce an object nothing on this // machine can link or run. - std::println(std::cerr, - "mcpp.rules.hip: [build] accel names AMD architectures ({}) and no ROCm\n" + std::cerr << std::format("mcpp.rules.hip: [build] accel names AMD architectures ({}) and no ROCm\n" " payload is published in this ecosystem yet, so nothing could link or run\n" " the result. The NVIDIA platform is available today:\n" " accel = \"hip, cuda12.9+{{sm_89}}\"\n" " which reaches the device through the CUDA runtime, with HIP as the API.", - tg.amd_archs.empty() ? std::string("none") : tg.amd_archs.front()); + tg.amd_archs.empty() ? std::string("none") : tg.amd_archs.front()) << '\n'; return out; } if (tg.cuda_archs.empty()) { // A device build that names no device is refused here, not at run time // as `no kernel image is available for execution`. - std::println(std::cerr, - "mcpp.rules.hip: [build] accel names no device architecture (accel = \"{}\").\n" + std::cerr << std::format("mcpp.rules.hip: [build] accel names no device architecture (accel = \"{}\").\n" " On the NVIDIA platform HIP compiles through the CUDA back end, and the\n" " device is spelled the way every other rule in this ecosystem spells it:\n" " accel = \"hip, cuda12.9+{{sm_89}}\"\n" " The set a build compiles for is a decision; the machine's own hardware is\n" " a poor default for it.", - mcpp::accel()); + mcpp::accel()) << '\n'; return out; } @@ -289,8 +305,7 @@ inline std::vector plan(std::span sources, options opt if (n.root->empty()) missing += std::format(" \"xim:{}\" = \"{}\"\n", n.pkg, n.version); if (!missing.empty()) { - std::println(std::cerr, - "mcpp.rules.hip: the HIP island needs payloads that are not installed.\n" + std::cerr << std::format("mcpp.rules.hip: the HIP island needs payloads that are not installed.\n" " This rule DECLARES them, so a project normally writes nothing. Check, in " "order:\n" " mcpp older than 2026.9.6.6; `features = [\"rules-hip\"]` missing from the\n" @@ -298,7 +313,7 @@ inline std::vector plan(std::span sources, options opt " To pin different versions, name them in your own project and they win:\n\n" " [target.'cfg(accelerator = \"hip\")'.xlings.workspace]\n{}\n" " They are PAYLOADS: the version is the project's choice, not the machine's.", - missing); + missing) << '\n'; return out; } @@ -313,12 +328,21 @@ inline std::vector plan(std::span sources, options opt // everywhere else. `mcpp.rules.cuda` takes the same path for the same // reason. const std::string tcdir = mcpp::toolchain_dir(); + // The suffix is the host's. This lane reaches only Linux today -- the + // NVIDIA-platform header package is published for it alone -- so the + // Windows spelling is not exercised by anything. It is written anyway, + // because the alternative is a path that is wrong on a host this rule + // will one day be asked about, and a wrong path reports itself as a + // missing toolchain. +#if defined(_WIN32) + const std::string cc = tcdir + "/bin/clang++.exe"; +#else const std::string cc = tcdir + "/bin/clang++"; +#endif if (tcdir.empty() || !std::filesystem::exists(cc)) { - std::println(std::cerr, - "mcpp.rules.hip: the NVIDIA platform compiles through clang, and this " + std::cerr << std::format("mcpp.rules.hip: the NVIDIA platform compiles through clang, and this " "project's\n toolchain has no clang++ at {}.\n" - " Select an LLVM toolchain: [toolchain] default = \"llvm@22.1.8\"", cc); + " Select an LLVM toolchain: [toolchain] default = \"llvm@22.1.8\"", cc) << '\n'; return out; } @@ -382,9 +406,9 @@ inline std::vector plan(std::span sources, options opt // rule resolved the payload, so the rule names where its libraries are. for (auto const& d : tk.lib_dirs()) mcpp::link_search(d.c_str()); - std::println("mcpp.rules.hip: NVIDIA platform -- HIP {} over CUDA {}, {} for {}", + std::cout << std::format("mcpp.rules.hip: NVIDIA platform -- HIP {} over CUDA {}, {} for {}", hip_version(tk.hip_root), tg.cuda_version.empty() ? "?" : tg.cuda_version, - std::filesystem::path(cc).filename().string(), tg.cuda_archs.front()); + std::filesystem::path(cc).filename().string(), tg.cuda_archs.front()) << '\n'; for (auto const& src : sources) { const auto stem = std::filesystem::path(src).stem().string(); diff --git a/rules/spirv.cppm b/rules/spirv.cppm index 78b3b3b..865884f 100644 --- a/rules/spirv.cppm +++ b/rules/spirv.cppm @@ -58,6 +58,25 @@ export module mcpp.rules.spirv; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::rules::spirv { struct options { @@ -173,17 +192,43 @@ inline flavour classify(const std::string& path) { return flavour::none; } +// HOW THIS HOST SPELLS A PROGRAM, decided where the build program is compiled. +// +// The build program runs on the machine doing the building, so these are +// properties of the HOST and not of the target being compiled for -- a cross +// build from Linux to Windows still looks for `glslc`, because that is the +// binary about to be executed. +#if defined(_WIN32) +inline constexpr std::string_view kExeSuffix = ".exe"; +inline constexpr char kPathSep = ';'; +#else +inline constexpr std::string_view kExeSuffix = ""; +inline constexpr char kPathSep = ':'; +#endif + +// The first of `/` and `/.exe` that exists. Both are +// tried on every host rather than only the one whose suffix matches: a payload +// repacked with the other convention is then found instead of silently missed, +// and the cost is one `stat`. +inline std::string program_in(const std::filesystem::path& dir, std::string_view name) { + std::string bare(name); + for (auto const& n : { bare, bare + std::string(kExeSuffix) }) { + auto p = (dir / n).string(); + if (is_file(p)) return p; + } + return {}; +} + inline std::string first_on_path(const char* exe) { const char* path = std::getenv("PATH"); if (!path || !*path) return {}; std::string_view sv(path); for (std::size_t i = 0; i <= sv.size();) { - auto sep = sv.find(':', i); + auto sep = sv.find(kPathSep, i); auto dir = sv.substr(i, sep == std::string_view::npos ? sv.size() - i : sep - i); i = sep == std::string_view::npos ? sv.size() + 1 : sep + 1; if (dir.empty()) continue; - auto p = (std::filesystem::path(dir) / exe).string(); - if (is_file(p)) return p; + if (auto p = program_in(std::filesystem::path(dir), exe); !p.empty()) return p; } return {}; } @@ -200,10 +245,9 @@ inline compiler find_compiler(const options& opt) { // Named but unrecognised: taking it as glslang would pass glslang's // flags to something that is not glslang, and the error would name // a flag rather than this decision. - std::println(stderr, - "mcpp.rules.spirv: options::compiler names '{}', which is neither glslang\n" + std::cerr << std::format("mcpp.rules.spirv: options::compiler names '{}', which is neither glslang\n" " nor glslc by program name, and the two share almost no flags. Rename the\n" - " program or point at the real one.", opt.compiler); + " program or point at the real one.", opt.compiler) << '\n'; return { .reported = true }; } return { opt.compiler, k }; @@ -213,10 +257,10 @@ inline compiler find_compiler(const options& opt) { if (const char* dir = mcpp::xpkg_dir("glslang"); dir && *dir) for (const char* exe : {"glslangValidator", "glslang"}) - if (auto p = (std::filesystem::path(dir) / "bin" / exe).string(); is_file(p)) + if (auto p = program_in(std::filesystem::path(dir) / "bin", exe); !p.empty()) return { p, flavour::glslang }; if (const char* dir = mcpp::xpkg_dir("shaderc"); dir && *dir) - if (auto p = (std::filesystem::path(dir) / "bin" / "glslc").string(); is_file(p)) + if (auto p = program_in(std::filesystem::path(dir) / "bin", "glslc"); !p.empty()) return { p, flavour::glslc }; for (const char* exe : {"glslangValidator", "glslang"}) @@ -229,13 +273,38 @@ inline compiler find_compiler(const options& opt) { // magic, the second is the release. The release is what a floor compares, and // stating it as a fact is what makes a build log answer "which compiler // produced this SPIR-V" without anyone having to reproduce the build. +// `popen` is POSIX and Windows spells it `_popen`; the null device differs +// too. Both are named here so the call sites below read the same on every +// host -- the alternative is a `#if` around each one, and the one that gets +// forgotten is the one nobody compiles. +inline FILE* open_pipe(const std::string& cmd) { +#if defined(_WIN32) + return ::_popen(cmd.c_str(), "r"); +#else + return ::popen(cmd.c_str(), "r"); +#endif +} +inline void close_pipe(FILE* p) { +#if defined(_WIN32) + ::_pclose(p); +#else + ::pclose(p); +#endif +} +inline constexpr const char* kNullDevice = +#if defined(_WIN32) + "NUL"; +#else + "/dev/null"; +#endif + inline std::string run_and_capture(const std::string& cmd) { - FILE* p = ::popen(cmd.c_str(), "r"); + FILE* p = open_pipe(cmd); if (!p) return {}; std::string text; char buf[512]; while (std::fgets(buf, sizeof buf, p)) text += buf; - ::pclose(p); + close_pipe(p); return text; } @@ -257,7 +326,8 @@ inline bool has_optimizer(const std::string& exe) { } inline std::string compiler_version(const compiler& cc) { - const std::string text = run_and_capture("\"" + cc.path + "\" --version 2>/dev/null"); + const std::string text = run_and_capture("\"" + cc.path + "\" --version 2>" + + std::string(kNullDevice)); // glslc: `shaderc v2026.3 2fbab05...` on the first line. glslang: // `Glslang Version: 11:15.1.0`, whose first field is the SPIR-V generator // magic and whose second is the release. @@ -381,7 +451,7 @@ inline bool wrap_glslc_output(const std::string& header, const std::string& inc, const std::string& sym) { std::ofstream out{header, std::ios::trunc}; if (!out) { - std::println(stderr, "mcpp.rules.spirv: cannot write {}", header); + std::cerr << std::format("mcpp.rules.spirv: cannot write {}", header) << '\n'; return false; } out << "// Generated by mcpp.rules.spirv. glslc emits an initialiser list;\n" @@ -400,8 +470,7 @@ inline bool compile(std::span shaders, options opt = {}) { const auto cc = find_compiler(opt); if (!cc) { if (cc.reported) return false; - std::println(stderr, - "mcpp.rules.spirv: no shader compiler found.\n" + std::cerr << std::format("mcpp.rules.spirv: no shader compiler found.\n" " This rule DECLARES glslang, so a project normally writes nothing. Check, in " "order:\n" " mcpp older than 2026.9.6.6; `features = [\"rules-spirv\"]` missing from the\n" @@ -412,7 +481,7 @@ inline bool compile(std::span shaders, options opt = {}) { " \"xim:glslang\" = \"15.1.0\" # glslangValidator\n" " \"xim:shaderc\" = \"2026.3\" # glslc\n" "or name it: MCPP_GLSLANG=/path/to/glslangValidator, MCPP_GLSLC=/path/to/glslc,\n" - "or set options::compiler."); + "or set options::compiler.") << '\n'; return false; } // The fact is keyed on the flavour, not on a shared name: which of the two @@ -441,14 +510,55 @@ inline bool compile(std::span shaders, options opt = {}) { std::error_code ec; std::filesystem::create_directories(gen, ec); + // TWO SHADERS THAT DIFFER ONLY BY DIRECTORY PRODUCE ONE HEADER AND ONE + // SYMBOL, AND THAT HAS TO BE REFUSED HERE. + // + // The output name is the stem and the stage, as this rule documents, so + // `shaders/ui/text.vert` and `shaders/world/text.vert` both resolve to + // `text_vert.h` declaring `text_vert_spv`. Disambiguating by directory is + // not the fix: the SYMBOL would still collide the moment both headers + // reached one translation unit, and the naming rule is what consumers write + // `#include` lines against. + // + // Measured before this check existed: ninja caught it -- `multiple rules + // generate .../text_vert.h` -- so it was never silent. What it did not do + // is name the two SHADERS, say which rule produced them, or state the way + // out; and it arrives as a graph-loading failure rather than as this rule's + // refusal. A project with one shader per stage never meets it, which is why + // it survived: a graphics project organising shaders by purpose is the + // first to have two. + { + std::map seen; // output stem -> first source + for (auto const& src : shaders) { + const std::filesystem::path p(src); + const auto stage = stage_of(p.extension().string()); + if (stage.empty()) continue; // reported below, per source + const auto key = p.stem().string() + "_" + std::string(stage); + auto [it, fresh] = seen.try_emplace(key, src); + if (!fresh) { + std::cerr << std::format("mcpp.rules.spirv: two shaders map to one output.\n" + " {}\n" + " {}\n" + " both produce `{}.h` declaring `{}`, because the name is the " + "shader's stem\n" + " and its stage -- the directory is not part of it, and could not " + "be: two\n" + " headers reaching one translation unit would still collide on the " + "symbol.\n" + " fix: rename one of them, or compile only one.", + it->second, src, key, symbol_of(p.stem().string(), stage)) << '\n'; + return false; + } + } + } + for (auto const& src : shaders) { const std::filesystem::path p(src); const auto stage = stage_of(p.extension().string()); if (stage.empty()) { - std::println(stderr, - "mcpp.rules.spirv: {} has no shader stage. Both compilers derive the stage " + std::cerr << std::format("mcpp.rules.spirv: {} has no shader stage. Both compilers derive the stage " "from the extension; rename it to one of .comp .vert .frag .geom .tesc " - ".tese .mesh .task .rgen .rint .rahit .rchit .rmiss .rcall", src); + ".tese .mesh .task .rgen .rint .rahit .rchit .rmiss .rcall", src) << '\n'; return false; } const auto sym = symbol_of(p.stem().string(), stage); diff --git a/rules/sycl.cppm b/rules/sycl.cppm index 422a32a..cd204d3 100644 --- a/rules/sycl.cppm +++ b/rules/sycl.cppm @@ -73,6 +73,25 @@ export module mcpp.rules.sycl; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::rules::sycl { struct options { @@ -97,6 +116,30 @@ struct options { // The device is spelled the way every other rule in this ecosystem spells it, // so `sm_89` does not acquire a second spelling because the source file says // `.sycl` instead of `.cu`. +// THE TWO PUBLISHED SYCL TOOLCHAINS DIFFER IN MORE THAN A FILE SUFFIX. +// +// Upstream publishes `sycl_linux.tar.gz` and `sycl_windows.tar.gz` from one +// tag, and the compiler is the same compiler. What differs is everything +// around it: the host half of a SYCL unit compiles against libstdc++ and glibc +// on Linux and against MSVC's standard library on Windows, so the three +// payloads that exist to keep the host's copies out of the search list +// (`xim:gcc`, `xim:glibc`, `xim:linux-headers`) have no counterpart there -- +// clang finds the MSVC installation itself, the same way it does for every +// ordinary translation unit on that host. `-fPIC` likewise names a property +// that is unconditional on Windows. +// +// And the device coverage differs: upstream states that the HIP and CUDA +// plugins are not built for Windows, and the asset agrees -- it carries +// Level Zero and OpenCL adapters and no others. An ahead-of-time NVIDIA +// build is therefore refused there rather than attempted. +#if defined(_WIN32) +inline constexpr bool kWindows = true; +inline constexpr const char* kExe = ".exe"; +#else +inline constexpr bool kWindows = false; +inline constexpr const char* kExe = ""; +#endif + struct target { bool sycl = false; std::vector cuda_archs; // {"sm_89"} @@ -174,13 +217,29 @@ inline std::string gcc_install_dir(const std::string& gcc_root) { // `dpcpp --version` states the release and the intel/llvm revision it was // built from. Stated as a fact so a build log answers "which SYCL compiler" // without anyone reproducing the build. +inline FILE* open_pipe(const std::string& cmd) { +#if defined(_WIN32) + return ::_popen(cmd.c_str(), "r"); +#else + return ::popen(cmd.c_str(), "r"); +#endif +} +inline void close_pipe(FILE* p) { +#if defined(_WIN32) + ::_pclose(p); +#else + ::pclose(p); +#endif +} + inline std::string compiler_version(const std::string& exe) { - FILE* p = ::popen(("\"" + exe + "\" --version 2>/dev/null").c_str(), "r"); + FILE* p = open_pipe("\"" + exe + "\" --version 2>" + + (kWindows ? std::string("NUL") : std::string("/dev/null"))); if (!p) return {}; std::string text; char buf[512]; while (std::fgets(buf, sizeof buf, p)) text += buf; - ::pclose(p); + close_pipe(p); for (auto line : split(text, '\n')) { auto at = line.find("DPC++ compiler "); if (at == std::string_view::npos) continue; @@ -235,18 +294,27 @@ inline std::vector plan(std::span sources, options opt std::vector out; const std::string root = mcpp::manifest_dir(); if (root.empty()) { - std::println(std::cerr, - "mcpp.rules.sycl: no mcpp build context -- this runs from build.mcpp"); + std::cerr << std::format("mcpp.rules.sycl: no mcpp build context -- this runs from build.mcpp") << '\n'; return out; } const auto tg = parse_target(mcpp::accel()); + if constexpr (kWindows) { + if (!tg.cuda_archs.empty()) { + std::cerr << std::format("mcpp.rules.sycl: [build] accel names an NVIDIA target and this host's SYCL\n" + " compiler cannot reach it. Upstream states that the CUDA and HIP plugins are\n" + " not built for Windows, and the published asset agrees: its Unified Runtime\n" + " adapters are Level Zero and OpenCL, and no others.\n" + " Available on this host: accel = \"sycl\" -- SPIR-V, consumed by whichever\n" + " Level Zero or OpenCL device the runtime finds.") << '\n'; + return out; + } + } if (!tg.amd_archs.empty() && tg.cuda_archs.empty()) { - std::println(std::cerr, - "mcpp.rules.sycl: [build] accel names AMD architectures and this ecosystem\n" + std::cerr << std::format("mcpp.rules.sycl: [build] accel names AMD architectures and this ecosystem\n" " publishes no ROCm payload yet, so nothing could link or run the result.\n" " Available today: accel = \"sycl\" (SPIR-V, any device the runtime finds)\n" - " or accel = \"sycl, cuda12.9+{{sm_89}}\" (ahead of time for NVIDIA)."); + " or accel = \"sycl, cuda12.9+{{sm_89}}\" (ahead of time for NVIDIA).") << '\n'; return out; } @@ -282,6 +350,11 @@ inline std::vector plan(std::span sources, options opt // does not need it. Asking for both would tell someone who has already // solved this to solve it again. if (dpcpp.empty() && opt.compiler.empty()) missing += " \"xim:dpcpp\" = \"7.1.0\"\n"; + // THE THREE BELOW ARE THE HOST C AND C++ LIBRARIES, AND ONLY LINUX HAS + // THIS PROBLEM. Requiring them on Windows would refuse a build over three + // packages that this ecosystem does not publish for it and that the + // compiler there does not need -- an error whose remedy does not exist. + if constexpr (!kWindows) { if (gcc.empty()) missing += " \"xim:gcc\" = \"15.1.0\"\n"; // UNPINNED ON PURPOSE, and this is the one detail that makes the // declaration portable. The C library version is the RUNTIME BINDING's @@ -294,11 +367,11 @@ inline std::vector plan(std::span sources, options opt // truthfully say about a library it does not select. if (glibc.empty()) missing += " \"xim:glibc\" = \"\"\n"; if (uapi.empty()) missing += " \"xim:linux-headers\" = \"\"\n"; + } if (!tg.cuda_archs.empty() && cuda.empty()) missing += " \"xim:cuda-nvcc\" = \"12.9.86\"\n"; if (!missing.empty()) { - std::println(std::cerr, - "mcpp.rules.sycl: the SYCL island needs payloads that are not installed.\n" + std::cerr << std::format("mcpp.rules.sycl: the SYCL island needs payloads that are not installed.\n" " This rule DECLARES them, so a project normally writes nothing. Check, in " "order:\n" " mcpp older than 2026.9.6.6; `features = [\"rules-sycl\"]` missing from the\n" @@ -310,26 +383,27 @@ inline std::vector plan(std::span sources, options opt " library underneath it. Without them dpcpp's clang reads the HOST's headers,\n" " which is measurable in its include search list and invisible on its command\n" " line.", - missing); + missing) << '\n'; return out; } auto exe = opt.compiler; - if (exe.empty()) exe = dpcpp + "/bin/clang++"; + if (exe.empty()) exe = dpcpp + "/bin/clang++" + kExe; if (!is_file(exe)) { - std::println(std::cerr, - "mcpp.rules.sycl: {} is not a file. The dpcpp payload publishes its SYCL\n" - " compiler under clang's own name; set options::compiler to name another.", exe); + std::cerr << std::format("mcpp.rules.sycl: {} is not a file. The dpcpp payload publishes its SYCL\n" + " compiler under clang's own name; set options::compiler to name another.", exe) << '\n'; return out; } if (auto v = compiler_version(exe); !v.empty()) mcpp::fact("dpcpp", v.c_str()); - const auto gid = gcc_install_dir(gcc); - if (gid.empty()) { - std::println(std::cerr, - "mcpp.rules.sycl: the xim:gcc payload at {} has no lib/gcc//\n" - " directory, which is what --gcc-install-dir names.", gcc); - return out; + // Empty on Windows, where the flag it feeds is not passed at all. + const auto gid = kWindows ? std::string{} : gcc_install_dir(gcc); + if constexpr (!kWindows) { + if (gid.empty()) { + std::cerr << std::format("mcpp.rules.sycl: the xim:gcc payload at {} has no lib/gcc//\n" + " directory, which is what --gcc-install-dir names.", gcc) << '\n'; + return out; + } } // The target selection, once, shared by the compile and the device link: @@ -344,23 +418,28 @@ inline std::vector plan(std::span sources, options opt } } - std::vector front{ exe, "-fsycl", "-std=c++17", "-O2", "-fPIC", - "--gcc-install-dir=" + gid }; - // The C library, ahead of whatever the compiler would have found. This is - // the shape mcpp uses for its own translation units, and it puts the - // ecosystem's glibc at the front of the search list; `/usr/include` stays - // last, as a fallback for C headers no payload provides, which is what the - // engine does too. - front.push_back("-isystem" + glibc + "/include"); - front.push_back("-isystem" + uapi + "/include"); + std::vector front{ exe, "-fsycl", "-std=c++17", "-O2" }; + if constexpr (!kWindows) { + front.push_back("-fPIC"); + front.push_back("--gcc-install-dir=" + gid); + // The C library, ahead of whatever the compiler would have found. This + // is the shape mcpp uses for its own translation units, and it puts + // the ecosystem's glibc at the front of the search list; + // `/usr/include` stays last, as a fallback for C headers no payload + // provides, which is what the engine does too. + front.push_back("-isystem" + glibc + "/include"); + front.push_back("-isystem" + uapi + "/include"); + } front.insert(front.end(), targeting.begin(), targeting.end()); // The link line gets its directories from here, not from the manifest: the // rule resolved the payload, so the rule names where its libraries are. mcpp::link_search((dpcpp + "/lib").c_str()); mcpp::link_lib("sycl"); - // See the file header for why this is not `-lstdc++`. - mcpp::link_lib(":libstdc++.so.6"); + // See the file header for why this is not `-lstdc++`. The reason is a + // Linux one: two C++ runtimes cannot share a process, and on Windows there + // is one -- MSVC's, which both this payload and mcpp's own compiler use. + if constexpr (!kWindows) mcpp::link_lib(":libstdc++.so.6"); // SPIR-V IS NOT A DEVICE, AND A BUILD THAT NAMES NO DEVICE SHOULD BE TOLD. // @@ -386,11 +465,11 @@ inline std::vector plan(std::span sources, options opt "scheduler, where the program cannot catch it. Name the device to " "compile ahead of time: accel = \"sycl, cuda12.9+{sm_89}\"."); - std::println("mcpp.rules.sycl: {} -- {} for {}", + std::cout << std::format("mcpp.rules.sycl: {} -- {} for {}", tg.cuda_archs.empty() ? "SPIR-V, compiled by the runtime" : "ahead of time, NVIDIA back end", std::filesystem::path(exe).filename().string(), - tg.cuda_archs.empty() ? std::string("any device") : tg.cuda_archs.front()); + tg.cuda_archs.empty() ? std::string("any device") : tg.cuda_archs.front()) << '\n'; std::vector objects; for (auto const& src : sources) { @@ -423,7 +502,11 @@ inline std::vector plan(std::span sources, options opt d.id = "sycl:device-link"; d.role = "object"; d.description = "dpcpp -fsycl-link (device images -> registration)"; - d.command = { exe, "-fsycl", "-fPIC", "--gcc-install-dir=" + gid }; + d.command = { exe, "-fsycl" }; + if constexpr (!kWindows) { + d.command.push_back("-fPIC"); + d.command.push_back("--gcc-install-dir=" + gid); + } d.command.insert(d.command.end(), targeting.begin(), targeting.end()); d.command.push_back("-fsycl-link"); for (auto const& o : objects) d.command.push_back(o); diff --git a/tests/all-rules-compile/build.mcpp b/tests/all-rules-compile/build.mcpp new file mode 100644 index 0000000..f265633 --- /dev/null +++ b/tests/all-rules-compile/build.mcpp @@ -0,0 +1,31 @@ +// Imports every rule and calls it. The call matters as much as the import: a +// module can be well-formed and still hold a template or an `if constexpr` +// branch that is only instantiated at a call site, and a rule that refuses to +// build the host it is being compiled for should say so here rather than in +// someone's project. +import std; +import mcpp; +import mcpp.rules.ascendc; +import mcpp.rules.cuda; +import mcpp.rules.hip; +import mcpp.rules.spirv; +import mcpp.rules.sycl; +import mcpp.tools.embed; + +int main() { + // No accelerator is named, so each of these returns true without looking + // for a payload. What is asserted is that the module compiled and the + // entry point resolved on this host. + bool ok = mcpp::rules::ascendc::compile() + && mcpp::rules::cuda::compile() + && mcpp::rules::hip::compile() + && mcpp::rules::spirv::compile() + && mcpp::rules::sycl::compile(); + // `tools::embed` has no accelerator gate, so it is asked the one question + // that writes nothing: where it WOULD put a header. `embed::file` is not + // called, because a fixture whose point is a compilation should not also + // produce output. + ok = ok && !mcpp::tools::embed::header_path("probe.bin").empty(); + std::println("all-rules-compile: every rule module compiled for this host"); + return ok ? 0 : 1; +} diff --git a/tests/all-rules-compile/mcpp.toml b/tests/all-rules-compile/mcpp.toml new file mode 100644 index 0000000..4c3f3e4 --- /dev/null +++ b/tests/all-rules-compile/mcpp.toml @@ -0,0 +1,51 @@ +# The fixture that compiles every rule in this collection, on whatever host is +# running the build. +# +# WHY A FIXTURE WHOSE ONLY PRODUCT IS A COMPILATION. +# +# Each of the other consumers exercises one rule end to end, and each of them +# needs that rule's payload: a shader compiler, a CUDA toolkit, a SYCL +# compiler. Those payloads are published for one, two or three platforms, so +# the consumers that need them can only run where they exist -- and the halves +# of a rule that are written for a host are exactly the halves that host never +# compiles. +# +# A rule's host-dependent code is `#if defined(_WIN32)`, a `.exe` suffix, a +# directory named `lib/x64`, a listing command. None of it is compiled by a +# Linux build, so a Windows branch can be syntactically wrong for as long as +# nobody builds on Windows -- and the first person to do so gets a compile +# error out of a package they only wanted to use. +# +# This fixture names no accelerator. Every rule's `compile()` returns +# immediately in that state, so no payload is installed and no second compiler +# is needed; what happens is that all six modules are compiled as host modules +# for this host, which is the assertion. It is cheap enough to run on every +# platform in the matrix. +[package] +name = "all-rules-compile" +namespace = "example" +version = "0.1.0" +description = "Compiles every rule module in this collection for the host running the build" + +[language] +standard = "c++23" +modules = true +import_std = true + +# Every feature this package publishes. A rule left out of this list is a rule +# whose host-dependent code is compiled on one platform only. +[build-dependencies.mcpp] +plugins = { path = "../..", features = [ + "rules-ascendc", "rules-cuda", "rules-hip", "rules-spirv", "rules-sycl", + "tools-embed", +], host-module = true } + +# NO `accel`, and that is the whole design: with none, every rule returns +# before it looks for a payload, so this fixture downloads nothing on any host. + +[build] +sources = ["src/*.cpp"] + +[targets.all-rules-compile] +kind = "bin" +main = "src/main.cpp" diff --git a/tests/all-rules-compile/src/main.cpp b/tests/all-rules-compile/src/main.cpp new file mode 100644 index 0000000..84b5a8c --- /dev/null +++ b/tests/all-rules-compile/src/main.cpp @@ -0,0 +1,9 @@ +#include + +// The program exists so the fixture has something to link. What is being +// tested happened before it: the build program compiled every rule module in +// the collection for this host. +int main() { + std::puts("all-rules-compile ok"); + return 0; +} diff --git a/tools/embed.cppm b/tools/embed.cppm index dad7fb2..6a1dd82 100644 --- a/tools/embed.cppm +++ b/tools/embed.cppm @@ -32,6 +32,25 @@ export module mcpp.tools.embed; import std; import mcpp; + +// WHY NOTHING HERE USES `std::println`, AND WHY THAT IS NOT A STYLE CHOICE. +// +// `std::print` and `std::println` are not header-only. Both of their overloads +// reach into the libc++ DYLIB -- `__is_posix_terminal(FILE*)` for the stdout +// form and `__get_ostream_file(ostream&)` for the stream form -- and those +// symbols were added to that library in a version macOS 14 does not ship. A +// build program's link resolves `-lc++` to the system copy there, so a rule +// that printed with `std::println` compiled and then failed to link: +// +// ld64.lld: error: undefined symbol: std::__1::__is_posix_terminal(__sFILE*) +// +// naming neither the call that needed it nor the reason. Measured on +// macos-14; macos-15 has the symbol, which is why nothing saw this until a +// rule was first compiled on the older of the two supported releases. +// +// `std::format` is header-only and has no such dependency, so every message in +// this file is formatted and then streamed. + export namespace mcpp::tools::embed { // The element the array is made of. A byte array is the general answer; a @@ -116,15 +135,14 @@ inline bool file(const std::filesystem::path& input, options opt = {}) { std::ifstream in(absolute, std::ios::binary); if (!in) { - std::println(stderr, "mcpp.tools.embed: cannot read {}", absolute.string()); + std::cerr << std::format("mcpp.tools.embed: cannot read {}", absolute.string()) << '\n'; return false; } std::string bytes((std::istreambuf_iterator(in)), std::istreambuf_iterator()); if (opt.elem == element::word32 && bytes.size() % 4 != 0) { - std::println(stderr, - "mcpp.tools.embed: {} is {} bytes, which is not a multiple of 4, and " - "element::word32 was asked for", absolute.string(), bytes.size()); + std::cerr << std::format("mcpp.tools.embed: {} is {} bytes, which is not a multiple of 4, and " + "element::word32 was asked for", absolute.string(), bytes.size()) << '\n'; return false; } @@ -171,7 +189,7 @@ inline bool file(const std::filesystem::path& input, options opt = {}) { if (!opt.name_space.empty()) text += "\n} // namespace " + opt.name_space + "\n"; if (!write_if_different(out, text)) { - std::println(stderr, "mcpp.tools.embed: cannot write {}", out.string()); + std::cerr << std::format("mcpp.tools.embed: cannot write {}", out.string()) << '\n'; return false; } @@ -188,8 +206,8 @@ inline bool file(const std::filesystem::path& input, options opt = {}) { // to the first input only. inline bool files(std::span inputs, options opt = {}) { if (!opt.identifier.empty()) { - std::println(stderr, "mcpp.tools.embed: options::identifier names one " - "symbol and files() writes several; call file() per input"); + std::cerr << std::format("mcpp.tools.embed: options::identifier names one " + "symbol and files() writes several; call file() per input") << '\n'; return false; } for (auto const& one : inputs)