From f263078e24f2e66264a1913ff4d71393a82d76b8 Mon Sep 17 00:00:00 2001 From: Luna Nova Date: Wed, 8 Oct 2025 08:16:12 -0700 Subject: [PATCH] rocmPackages.hipblaslt: stop inlining war and peace in asm comments Reduces peak build dir usage to 150G hipblaslt was filling the build tmpfs, peaking at 240G in the build/Tensile/build_tmp/TENSILEdir with this bodge we stop it from writing very verbose comments in the generated assembly kernels example excessive whitespace/comment content: s_nop 0 // 1 wait state required when next inst writes vgprs held by previous dwordx4 store inst --- .../rocm-modules/6/hipblaslt/default.nix | 3 ++ .../6/hipblaslt/reduce-comment-spam.patch | 47 +++++++++++++++++++ 2 files changed, 50 insertions(+) create mode 100644 pkgs/development/rocm-modules/6/hipblaslt/reduce-comment-spam.patch diff --git a/pkgs/development/rocm-modules/6/hipblaslt/default.nix b/pkgs/development/rocm-modules/6/hipblaslt/default.nix index 47bfd00909fc9..effc0c2f97c3a 100644 --- a/pkgs/development/rocm-modules/6/hipblaslt/default.nix +++ b/pkgs/development/rocm-modules/6/hipblaslt/default.nix @@ -111,6 +111,9 @@ stdenv.mkDerivation (finalAttrs: { # Support loading zstd compressed .dat files, required to keep output under # hydra size limit ./messagepack-compression-support.patch + # excessive comments are written to temporary asm files in build dir + # TODO: report upstream, find a better solution + ./reduce-comment-spam.patch ]; postPatch = '' diff --git a/pkgs/development/rocm-modules/6/hipblaslt/reduce-comment-spam.patch b/pkgs/development/rocm-modules/6/hipblaslt/reduce-comment-spam.patch new file mode 100644 index 0000000000000..32969c5177810 --- /dev/null +++ b/pkgs/development/rocm-modules/6/hipblaslt/reduce-comment-spam.patch @@ -0,0 +1,47 @@ +diff --git a/projects/hipblaslt/tensilelite/rocisa/rocisa/include/format.hpp b/projects/hipblaslt/tensilelite/rocisa/rocisa/include/format.hpp +index b7dcb6f59a..b0625ba769 100644 +--- a/tensilelite/rocisa/rocisa/include/format.hpp ++++ b/tensilelite/rocisa/rocisa/include/format.hpp +@@ -8,11 +8,13 @@ namespace rocisa + // Text format functions + inline std::string slash(const std::string& comment) + { ++ return ""; + return "// " + comment + "\n"; + } + + inline std::string slash50(const std::string& comment) + { ++ return ""; + std::ostringstream oss; + oss << std::setw(50) << "" + << " // " << comment << "\n"; +@@ -21,16 +23,19 @@ namespace rocisa + + inline std::string block(const std::string& comment) + { ++ return ""; + return "/* " + comment + " */\n"; + } + + inline std::string blockNewLine(const std::string& comment) + { ++ return ""; + return "\n/* " + comment + " */\n"; + } + + inline std::string block3Line(const std::string& comment) + { ++ return ""; + std::ostringstream oss; + oss << "\n/******************************************/\n"; + std::istringstream iss(comment); +@@ -52,7 +57,7 @@ namespace rocisa + { + formattedStr = "\"" + formattedStr + "\\n\\t\""; + } +- if(!comment.empty()) ++ if(false) + { + std::string buffer = formattedStr + + std::string(std::max(0, 50 - int(formattedStr.length())), ' ')