From 1be3d3c8fe73d7264605256fabc5a37fe797d49b Mon Sep 17 00:00:00 2001 From: Gaetan Lepage Date: Mon, 12 Jan 2026 17:06:42 +0000 Subject: [PATCH 1/2] python3Packages.tinygrad: 0.11.0 -> 0.12.0 Diff: https://github.com/tinygrad/tinygrad/compare/v0.11.0...v0.12.0 Changelog: https://github.com/tinygrad/tinygrad/releases/tag/v0.12.0 --- .../python-modules/tinygrad/default.nix | 252 ++++++++++-------- .../tinygrad/fix-dlopen-cuda.patch | 49 ---- .../tinygrad/patch-cuda-paths.patch | 39 +++ .../tinygrad/patch-deps-paths.patch | 39 +++ .../tinygrad/patch-rocm-paths.patch | 68 +++++ 5 files changed, 287 insertions(+), 160 deletions(-) delete mode 100644 pkgs/development/python-modules/tinygrad/fix-dlopen-cuda.patch create mode 100644 pkgs/development/python-modules/tinygrad/patch-cuda-paths.patch create mode 100644 pkgs/development/python-modules/tinygrad/patch-deps-paths.patch create mode 100644 pkgs/development/python-modules/tinygrad/patch-rocm-paths.patch diff --git a/pkgs/development/python-modules/tinygrad/default.nix b/pkgs/development/python-modules/tinygrad/default.nix index 0eaa076fc091..cf2992f7bc76 100644 --- a/pkgs/development/python-modules/tinygrad/default.nix +++ b/pkgs/development/python-modules/tinygrad/default.nix @@ -17,35 +17,43 @@ setuptools, # optional-dependencies - llvmlite, - triton, + # arm unicorn, + # triton + triton, + # testing_minimal + hypothesis, + numpy, + pytest-xdist, + torch, + z3-solver, + # testing_unit + ggml-python, + openai, + safetensors, + tabulate, + tqdm, + # testing + blobfile, + boto3, + bottle, + capstone, + librosa, + networkx, + nibabel, + onnx, + onnxruntime, + opencv4, + pandas, + pillow, + pycocotools, + sentencepiece, + tiktoken, + transformers, # tests pytestCheckHook, writableTmpDirAsHomeHook, - blobfile, - bottle, - capstone, - clang, - hexdump, - hypothesis, - jax, - librosa, - ml-dtypes, - networkx, - numpy, - onnx, - onnxruntime, - pillow, - pytest-xdist, - safetensors, - sentencepiece, - tiktoken, - torch, - tqdm, - transformers, - z3-solver, # passthru tinygrad, @@ -56,118 +64,124 @@ buildPythonPackage (finalAttrs: { pname = "tinygrad"; - version = "0.11.0"; + version = "0.12.0"; pyproject = true; src = fetchFromGitHub { owner = "tinygrad"; repo = "tinygrad"; tag = "v${finalAttrs.version}"; - hash = "sha256-VG2rhkiwPFN3JYSBbqrwCdqhdGE8GY6oEatMSCydhw8="; + hash = "sha256-Lied19C1sAbislr2WznnCZEmOn5PA0OzMg2KOdWOYkA="; }; - patches = [ - (replaceVars ./fix-dlopen-cuda.patch { - inherit (addDriverRunpath) driverLink; - libnvrtc = - if cudaSupport then - "${lib.getLib cudaPackages.cuda_nvrtc}/lib/libnvrtc.so" - else - "Please import nixpkgs with `config.cudaSupport = true`"; - }) - ]; + patches = + let + libExtension = stdenv.hostPlatform.extensions.sharedLibrary; + in + [ + (replaceVars ./patch-deps-paths.patch { + libllvm = "${lib.getLib llvmPackages.llvm}/lib/libLLVM${libExtension}"; + libclang = "${lib.getLib llvmPackages.libclang}/lib/libclang${libExtension}"; - postPatch = - # Patch `clang` directly in the source file - # Use the unwrapped variant to enable the "native" features currently unavailable in the sandbox - '' - substituteInPlace tinygrad/runtime/ops_cpu.py \ - --replace-fail "getenv(\"CC\", 'clang')" "'${lib.getExe llvmPackages.clang-unwrapped}'" - '' - + '' - substituteInPlace tinygrad/runtime/support/llvm.py \ - --replace-fail "ctypes.util.find_library('LLVM')" "'${lib.getLib llvmPackages.llvm}/lib/libLLVM.so'" - '' - + lib.optionalString stdenv.hostPlatform.isLinux '' - substituteInPlace tinygrad/runtime/autogen/libc.py \ - --replace-fail "ctypes.util.find_library('c')" "'${stdenv.cc.libc}/lib/libc.so.6'" + # Use the unwrapped variant to enable the "native" features currently unavailable in the sandbox + clang = lib.getExe llvmPackages.clang-unwrapped; + }) + ] + ++ lib.optionals cudaSupport [ + (replaceVars ./patch-cuda-paths.patch { + inherit (addDriverRunpath) driverLink; + cuda_nvrtc = lib.getLib cudaPackages.cuda_nvrtc; - substituteInPlace tinygrad/runtime/autogen/opencl.py \ - --replace-fail "ctypes.util.find_library('OpenCL')" "'${ocl-icd}/lib/libOpenCL.so'" - '' - # test/test_tensor.py imports the PTX variable from the cuda_compiler.py file. - # This import leads to loading the libnvrtc.so library that is not substituted when cudaSupport = false. - # -> As a fix, we hardcode this variable to False - + lib.optionalString (!cudaSupport) '' - substituteInPlace test/test_tensor.py \ - --replace-fail "from tinygrad.runtime.support.compiler_cuda import PTX" "PTX = False" - '' - # `cuda_fp16.h` and co. are needed at runtime to compile kernels - + lib.optionalString cudaSupport '' - substituteInPlace tinygrad/runtime/support/compiler_cuda.py \ - --replace-fail \ - '"-I/usr/local/cuda/include", "-I/usr/include", "-I/opt/cuda/include"' \ - '"-I${lib.getDev cudaPackages.cuda_cudart}/include/"' - '' - + lib.optionalString rocmSupport '' - substituteInPlace tinygrad/runtime/autogen/hip.py \ - --replace-fail "/opt/rocm/" "${rocmPackages.clr}/" + # `cuda_fp16.h` and co. are needed at runtime to compile kernels + cuda_cudart = lib.getInclude cudaPackages.cuda_cudart; + }) + ] + ++ lib.optionals rocmSupport [ + (replaceVars ./patch-rocm-paths.patch { + comgr = lib.getLib rocmPackages.rocm-comgr; + clr = lib.getLib rocmPackages.clr; + rocm-runtime = lib.getLib rocmPackages.rocm-runtime; + rocm-llvm-objdump = lib.getExe' rocmPackages.llvm.llvm "llvm-objdump"; + llvm-objdump = lib.getExe' llvmPackages.llvm "llvm-objdump"; + hipcc = lib.getExe' rocmPackages.hipcc "hipcc"; + }) + ]; - substituteInPlace tinygrad/runtime/support/compiler_hip.py \ - --replace-fail "/opt/rocm/include" "${rocmPackages.clr}/include" + postPatch = lib.optionalString stdenv.hostPlatform.isLinux '' + substituteInPlace tinygrad/runtime/autogen/opencl.py \ + --replace-fail \ + "dll = DLL('opencl', 'OpenCL')" \ + "dll = DLL('opencl', '${lib.getLib ocl-icd}/lib/libOpenCL.so')" - substituteInPlace tinygrad/runtime/support/compiler_hip.py \ - --replace-fail "/opt/rocm/llvm" "${rocmPackages.llvm.llvm}" - - substituteInPlace tinygrad/runtime/autogen/comgr.py \ - --replace-fail "/opt/rocm/" "${rocmPackages.rocm-comgr}/" - ''; + substituteInPlace tinygrad/runtime/autogen/libc.py \ + --replace-fail \ + "dll = DLL('libc', 'c', use_errno=True)" \ + "dll = DLL('libc', '${lib.getLib stdenv.cc.libc}/lib/libc.so.6', use_errno=True)" + ''; __propagatedImpureHostDeps = lib.optional stdenv.hostPlatform.isDarwin "/usr/lib/libc.dylib"; build-system = [ setuptools ]; - optional-dependencies = { - llvm = [ llvmlite ]; + optional-dependencies = lib.fix (self: { arm = [ unicorn ]; triton = [ triton ]; - }; + testing_minimal = [ + hypothesis + numpy + pytest-xdist + torch + z3-solver + ]; + testing_unit = self.testing_minimal ++ [ + ggml-python + openai + safetensors + tabulate + tqdm + ]; + testing = self.testing_unit ++ [ + blobfile + boto3 + bottle + capstone + librosa + networkx + nibabel + onnx + onnxruntime + opencv4 + pandas + pillow + pycocotools + sentencepiece + tiktoken + transformers + ]; + }); pythonImportsCheck = [ "tinygrad" + "tinygrad.runtime.autogen.libclang" ] ++ lib.optionals cudaSupport [ + "tinygrad.runtime.ops_cuda" "tinygrad.runtime.ops_nv" + ] + ++ lib.optionals rocmSupport [ + "tinygrad.runtime.ops_amd" + "tinygrad.runtime.ops_hip" + ] + ++ lib.optionals stdenv.hostPlatform.isDarwin [ + "tinygrad.runtime.ops_metal" ]; nativeCheckInputs = [ + llvmPackages.clang pytestCheckHook writableTmpDirAsHomeHook - - blobfile - bottle - capstone - clang - hexdump - hypothesis - jax - librosa - ml-dtypes - networkx - numpy - onnx - onnxruntime - pillow - pytest-xdist - safetensors - sentencepiece - tiktoken - torch - tqdm - transformers - z3-solver ] - ++ networkx.optional-dependencies.extra; + ++ finalAttrs.passthru.optional-dependencies.testing; disabledTests = [ # RuntimeError: Attempting to relocate against an undefined symbol 'fmaxf' @@ -192,6 +206,7 @@ buildPythonPackage (finalAttrs: { "test_dataset_is_realized" "test_e2e_big" "test_fetch_small" + "test_hevc_parser" "test_huggingface_enet_safetensors" "test_index_mnist" "test_linear_mnist" @@ -206,8 +221,13 @@ buildPythonPackage (finalAttrs: { "test_load_convnext" "test_load_enet" "test_load_enet_alt" + "test_load_gpt2_q4_1" "test_load_llama2bfloat" "test_load_resnet" + "test_load_sample_mxfp4" + "test_load_sample_q6_k" + "test_load_tinyllama_q4_0" + "test_load_tinyllama_q8_0" "test_mnist_val" "test_openpilot_model" "test_resnet" @@ -220,12 +240,19 @@ buildPythonPackage (finalAttrs: { "test_transcribe_long_no_batch" "test_vgg7" ] - ++ lib.optionals (stdenv.hostPlatform.system == "aarch64-linux") [ + ++ lib.optionals (stdenv.hostPlatform.isLinux && stdenv.hostPlatform.isAarch64) [ # Fail with AssertionError "test_casts_from" "test_casts_to" + "test_float_cast_to_unsigned_underflow" "test_int8" "test_int8_to_uint16_negative" + ] + ++ lib.optionals stdenv.hostPlatform.isDarwin [ + # Flaky (pass when running a smaller set of tests: tests/unit/*, but not within the full test suite) + # AttributeError: module 'tinygrad.runtime.autogen.libclang' has no attribute 'clang_parseTranslationUnit' + "test_gen_from_header" + "test_struct_ordering" ]; disabledTestPaths = [ @@ -236,10 +263,18 @@ buildPythonPackage (finalAttrs: { # Files under this directory are not considered as tests by upstream and should be skipped "extra/" + ] + ++ lib.optionals stdenv.hostPlatform.isDarwin [ + # urllib.error.URLError: + "test/models/test_whisper.py" ]; + __darwinAllowLocalNetworking = true; + passthru.tests = { withCuda = tinygrad.override { cudaSupport = true; }; + withRocm = tinygrad.override { rocmSupport = true; }; }; meta = { @@ -248,10 +283,5 @@ buildPythonPackage (finalAttrs: { changelog = "https://github.com/tinygrad/tinygrad/releases/tag/${finalAttrs.src.tag}"; license = lib.licenses.mit; maintainers = with lib.maintainers; [ GaetanLepage ]; - badPlatforms = [ - # Fatal Python error: Aborted - # onnxruntime/capi/_pybind_state.py", line 32 in - "aarch64-linux" - ]; }; }) diff --git a/pkgs/development/python-modules/tinygrad/fix-dlopen-cuda.patch b/pkgs/development/python-modules/tinygrad/fix-dlopen-cuda.patch deleted file mode 100644 index 285ca6111569..000000000000 --- a/pkgs/development/python-modules/tinygrad/fix-dlopen-cuda.patch +++ /dev/null @@ -1,49 +0,0 @@ -diff --git a/tinygrad/runtime/autogen/cuda.py b/tinygrad/runtime/autogen/cuda.py -index a30c8f53..e2078ff6 100644 ---- a/tinygrad/runtime/autogen/cuda.py -+++ b/tinygrad/runtime/autogen/cuda.py -@@ -145,7 +145,19 @@ def char_pointer_cast(string, encoding='utf-8'): - - - _libraries = {} --_libraries['libcuda.so'] = ctypes.CDLL(ctypes.util.find_library('cuda')) -+libcuda = None -+try: -+ libcuda = ctypes.CDLL('libcuda.so') -+except OSError: -+ pass -+try: -+ libcuda = ctypes.CDLL('@driverLink@/lib/libcuda.so') -+except OSError: -+ pass -+if libcuda is None: -+ raise RuntimeError(f"`libcuda.so` not found") -+ -+_libraries['libcuda.so'] = libcuda - - - cuuint32_t = ctypes.c_uint32 -diff --git a/tinygrad/runtime/autogen/nvrtc.py b/tinygrad/runtime/autogen/nvrtc.py -index 6af74187..c5a6c6c4 100644 ---- a/tinygrad/runtime/autogen/nvrtc.py -+++ b/tinygrad/runtime/autogen/nvrtc.py -@@ -10,7 +10,18 @@ import ctypes, ctypes.util - - - _libraries = {} --_libraries['libnvrtc.so'] = ctypes.CDLL(ctypes.util.find_library('nvrtc')) -+libnvrtc = None -+try: -+ libnvrtc = ctypes.CDLL('libnvrtc.so') -+except OSError: -+ pass -+try: -+ libnvrtc = ctypes.CDLL('@libnvrtc@') -+except OSError: -+ pass -+if libnvrtc is None: -+ raise RuntimeError(f"`libnvrtc.so` not found") -+_libraries['libnvrtc.so'] = libnvrtc - def string_cast(char_pointer, encoding='utf-8', errors='strict'): - value = ctypes.cast(char_pointer, ctypes.c_char_p).value - if value is not None and encoding is not None: diff --git a/pkgs/development/python-modules/tinygrad/patch-cuda-paths.patch b/pkgs/development/python-modules/tinygrad/patch-cuda-paths.patch new file mode 100644 index 000000000000..b2a2177ab3e2 --- /dev/null +++ b/pkgs/development/python-modules/tinygrad/patch-cuda-paths.patch @@ -0,0 +1,39 @@ +diff --git a/tinygrad/runtime/autogen/cuda.py b/tinygrad/runtime/autogen/cuda.py +index 061c8d73f..65ec3ffd7 100644 +--- a/tinygrad/runtime/autogen/cuda.py ++++ b/tinygrad/runtime/autogen/cuda.py +@@ -1,7 +1,7 @@ + # mypy: ignore-errors + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR +-dll = DLL('cuda', 'cuda') ++dll = DLL('cuda', '@driverLink@/lib/libcuda.so') + cuuint32_t = ctypes.c_uint32 + cuuint64_t = ctypes.c_uint64 + CUdeviceptr_v2 = ctypes.c_uint64 +diff --git a/tinygrad/runtime/autogen/nvrtc.py b/tinygrad/runtime/autogen/nvrtc.py +index 88085c45b..90518d403 100644 +--- a/tinygrad/runtime/autogen/nvrtc.py ++++ b/tinygrad/runtime/autogen/nvrtc.py +@@ -2,7 +2,7 @@ + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR + import sysconfig +-dll = DLL('nvrtc', 'nvrtc', f'/usr/local/cuda/targets/{sysconfig.get_config_vars().get("MULTIARCH", "").rsplit("-", 1)[0]}/lib') ++dll = DLL('nvrtc','@cuda_nvrtc@/lib/libnvrtc.so') + nvrtcResult = CEnum(ctypes.c_uint32) + NVRTC_SUCCESS = nvrtcResult.define('NVRTC_SUCCESS', 0) + NVRTC_ERROR_OUT_OF_MEMORY = nvrtcResult.define('NVRTC_ERROR_OUT_OF_MEMORY', 1) +diff --git a/tinygrad/runtime/support/compiler_cuda.py b/tinygrad/runtime/support/compiler_cuda.py +index 8f71a9255..fdbf01bad 100644 +--- a/tinygrad/runtime/support/compiler_cuda.py ++++ b/tinygrad/runtime/support/compiler_cuda.py +@@ -43,7 +43,7 @@ def cuda_disassemble(lib:bytes, arch:str): + class CUDACompiler(Compiler): + def __init__(self, arch:str, cache_key:str="cuda"): + self.arch, self.compile_options = arch, [f'--gpu-architecture={arch}'] +- self.compile_options += [f"-I{CUDA_PATH}/include"] if CUDA_PATH else ["-I/usr/local/cuda/include", "-I/usr/include", "-I/opt/cuda/include"] ++ self.compile_options += ["-I@cuda_cudart@/include"] + nvrtc_check(nvrtc.nvrtcVersion((nvrtcMajor := ctypes.c_int()), (nvrtcMinor := ctypes.c_int()))) + if (nvrtcMajor.value, nvrtcMinor.value) >= (12, 4): self.compile_options.append("--minimal") + super().__init__(f"compile_{cache_key}_{self.arch}") diff --git a/pkgs/development/python-modules/tinygrad/patch-deps-paths.patch b/pkgs/development/python-modules/tinygrad/patch-deps-paths.patch new file mode 100644 index 000000000000..8e8bbeec2f66 --- /dev/null +++ b/pkgs/development/python-modules/tinygrad/patch-deps-paths.patch @@ -0,0 +1,39 @@ +diff --git a/tinygrad/runtime/autogen/libclang.py b/tinygrad/runtime/autogen/libclang.py +index dbf22b6f5..4a5a097dc 100644 +--- a/tinygrad/runtime/autogen/libclang.py ++++ b/tinygrad/runtime/autogen/libclang.py +@@ -1,7 +1,7 @@ + # mypy: ignore-errors + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR +-dll = DLL('libclang', ['clang-20', 'clang']) ++dll = DLL('libclang', '@libclang@') + CXIndex = ctypes.c_void_p + class struct_CXTargetInfoImpl(Struct): pass + CXTargetInfo = ctypes.POINTER(struct_CXTargetInfoImpl) +diff --git a/tinygrad/runtime/autogen/llvm.py b/tinygrad/runtime/autogen/llvm.py +index 7d05224e7..54204f563 100644 +--- a/tinygrad/runtime/autogen/llvm.py ++++ b/tinygrad/runtime/autogen/llvm.py +@@ -2,7 +2,7 @@ + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR + from tinygrad.helpers import WIN, OSX +-dll = DLL('llvm', 'C:\\Program Files\\LLVM\\bin\\LLVM-C.dll' if WIN else '/opt/homebrew/opt/llvm@20/lib/libLLVM.dylib' if OSX else ['LLVM', 'LLVM-21', 'LLVM-20', 'LLVM-19', 'LLVM-18', 'LLVM-17', 'LLVM-16', 'LLVM-15', 'LLVM-14']) ++dll = DLL('llvm', '@libllvm@') + intmax_t = ctypes.c_int64 + try: (imaxabs:=dll.imaxabs).restype, imaxabs.argtypes = intmax_t, [intmax_t] + except AttributeError: pass +diff --git a/tinygrad/runtime/support/compiler_cpu.py b/tinygrad/runtime/support/compiler_cpu.py +index 8b11f3af8..d8190fac3 100644 +--- a/tinygrad/runtime/support/compiler_cpu.py ++++ b/tinygrad/runtime/support/compiler_cpu.py +@@ -15,7 +15,7 @@ class ClangJITCompiler(Compiler): + arch = {'x86_64': '-march=native', 'AMD64': '-march=native', 'riscv64': '-march=rv64g'}.get(platform.machine(), "-mcpu=native") + args = [arch, f'--target={target}-none-unknown-elf', '-O2', '-fPIC', '-ffreestanding', '-fno-math-errno', '-nostdlib', '-fno-ident'] + arch_args = ['-ffixed-x18'] if target == 'arm64' else [] +- obj = subprocess.check_output([getenv("CC", 'clang'), '-c', '-x', 'c', *args, *arch_args, '-', '-o', '-'], input=src.encode('utf-8')) ++ obj = subprocess.check_output(["@clang@", '-c', '-x', 'c', *args, *arch_args, '-', '-o', '-'], input=src.encode('utf-8')) + return jit_loader(obj) + + def disassemble(self, lib:bytes): return capstone_flatdump(lib) diff --git a/pkgs/development/python-modules/tinygrad/patch-rocm-paths.patch b/pkgs/development/python-modules/tinygrad/patch-rocm-paths.patch new file mode 100644 index 000000000000..5b96c6e906f1 --- /dev/null +++ b/pkgs/development/python-modules/tinygrad/patch-rocm-paths.patch @@ -0,0 +1,68 @@ +diff --git a/tinygrad/runtime/autogen/comgr.py b/tinygrad/runtime/autogen/comgr.py +index 9c642cef2..b42c7c9c8 100644 +--- a/tinygrad/runtime/autogen/comgr.py ++++ b/tinygrad/runtime/autogen/comgr.py +@@ -2,7 +2,7 @@ + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR + import os +-dll = DLL('comgr', [os.getenv('ROCM_PATH', '/opt/rocm')+'/lib/libamd_comgr.so', 'amd_comgr']) ++dll = DLL('comgr', '@comgr@/lib/libamd_comgr.so') + amd_comgr_status_s = CEnum(ctypes.c_uint32) + AMD_COMGR_STATUS_SUCCESS = amd_comgr_status_s.define('AMD_COMGR_STATUS_SUCCESS', 0) + AMD_COMGR_STATUS_ERROR = amd_comgr_status_s.define('AMD_COMGR_STATUS_ERROR', 1) +diff --git a/tinygrad/runtime/autogen/hip.py b/tinygrad/runtime/autogen/hip.py +index 1cccb1e1f..d57a4d90a 100644 +--- a/tinygrad/runtime/autogen/hip.py ++++ b/tinygrad/runtime/autogen/hip.py +@@ -2,7 +2,7 @@ + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR + import os +-dll = DLL('hip', os.getenv('ROCM_PATH', '/opt/rocm')+'/lib/libamdhip64.so') ++dll = DLL('hip', '@clr@/lib/libamdhip64.so') + hipError_t = CEnum(ctypes.c_uint32) + hipSuccess = hipError_t.define('hipSuccess', 0) + hipErrorInvalidValue = hipError_t.define('hipErrorInvalidValue', 1) +diff --git a/tinygrad/runtime/autogen/hsa.py b/tinygrad/runtime/autogen/hsa.py +index 16e568e21..9788d3705 100644 +--- a/tinygrad/runtime/autogen/hsa.py ++++ b/tinygrad/runtime/autogen/hsa.py +@@ -2,7 +2,7 @@ + import ctypes + from tinygrad.runtime.support.c import DLL, Struct, CEnum, _IO, _IOW, _IOR, _IOWR + import os +-dll = DLL('hsa', [os.getenv('ROCM_PATH', '/opt/rocm')+'/lib/libhsa-runtime64.so', 'hsa-runtime64']) ++dll = DLL('hsa', '@rocm-runtime@/lib/libhsa-runtime64.so') + enum_SQ_RSRC_BUF_TYPE = CEnum(ctypes.c_uint32) + SQ_RSRC_BUF = enum_SQ_RSRC_BUF_TYPE.define('SQ_RSRC_BUF', 0) + SQ_RSRC_BUF_RSVD_1 = enum_SQ_RSRC_BUF_TYPE.define('SQ_RSRC_BUF_RSVD_1', 1) +diff --git a/tinygrad/runtime/support/compiler_amd.py b/tinygrad/runtime/support/compiler_amd.py +index e97aaefcd..ef55838a3 100644 +--- a/tinygrad/runtime/support/compiler_amd.py ++++ b/tinygrad/runtime/support/compiler_amd.py +@@ -13,9 +13,9 @@ from tinygrad.runtime.support.compiler_cpu import LLVMCompiler + from tinygrad.helpers import OSX, to_char_p_p + + def _find_llvm_objdump(): +- if OSX: return '/opt/homebrew/opt/llvm/bin/llvm-objdump' ++ if OSX: return '@llvm-objdump@' + # Try ROCm path first, then versioned, then unversioned +- for p in ['/opt/rocm/llvm/bin/llvm-objdump', 'llvm-objdump-21', 'llvm-objdump-20', 'llvm-objdump']: ++ for p in ['@rocm-llvm-objdump@', '@llvm-objdump@']: + if shutil.which(p): return p + raise FileNotFoundError("llvm-objdump not found") + +@@ -107,9 +107,9 @@ class HIPCCCompiler(Compiler): + srcf.write(src.encode()) + srcf.flush() + +- subprocess.run(["hipcc", "-c", "-emit-llvm", "--cuda-device-only", "-O3", "-mcumode", +- f"--offload-arch={self.arch}", "-I/opt/rocm/include/hip", "-o", bcf.name, srcf.name] + self.extra_options, check=True) +- subprocess.run(["hipcc", "-target", "amdgcn-amd-amdhsa", f"-mcpu={self.arch}", ++ subprocess.run(["@hipcc@", "-c", "-emit-llvm", "--cuda-device-only", "-O3", "-mcumode", ++ f"--offload-arch={self.arch}", "-I@clr@/include/hip", "-o", bcf.name, srcf.name] + self.extra_options, check=True) ++ subprocess.run(["@hipcc@", "-target", "amdgcn-amd-amdhsa", f"-mcpu={self.arch}", + "-O3", "-mllvm", "-amdgpu-internalize-symbols", "-c", "-o", libf.name, bcf.name] + self.extra_options, check=True) + + return pathlib.Path(libf.name).read_bytes() From 42d9ac4971357c9c42d746e3de9658cd5bac29d5 Mon Sep 17 00:00:00 2001 From: Gaetan Lepage Date: Fri, 16 Jan 2026 11:23:20 +0100 Subject: [PATCH 2/2] python3Packages.tinygrad: add passthru.gpuCheck to runs tests on GPU --- .../python-modules/tinygrad/default.nix | 18 +++++++++++++++--- 1 file changed, 15 insertions(+), 3 deletions(-) diff --git a/pkgs/development/python-modules/tinygrad/default.nix b/pkgs/development/python-modules/tinygrad/default.nix index cf2992f7bc76..56c29e669db5 100644 --- a/pkgs/development/python-modules/tinygrad/default.nix +++ b/pkgs/development/python-modules/tinygrad/default.nix @@ -272,9 +272,21 @@ buildPythonPackage (finalAttrs: { __darwinAllowLocalNetworking = true; - passthru.tests = { - withCuda = tinygrad.override { cudaSupport = true; }; - withRocm = tinygrad.override { rocmSupport = true; }; + passthru = { + tests = { + withCuda = tinygrad.override { cudaSupport = true; }; + withRocm = tinygrad.override { rocmSupport = true; }; + }; + + gpuCheck = tinygrad.overridePythonAttrs (old: { + requiredSystemFeatures = [ "cuda" ]; + + pytestFlags = (old.pytestFlags or [ ]) ++ [ + # When running in parallel, with GPU support, some tests become flaky: + # RuntimeError: Wait timeout: 30000 ms! (the signal is not set to 153, but 151) + "--maxprocesses=1" + ]; + }); }; meta = {