https://github.com/cseslowpoke created https://github.com/llvm/llvm-project/pull/177577
Fix #172426 >From cebf3d0c38381e753e1d1662e3311cd2a67c36d6 Mon Sep 17 00:00:00 2001 From: cseslowpoke <[email protected]> Date: Fri, 23 Jan 2026 12:54:17 +0100 Subject: [PATCH] [Clang][CUDA][Driver] Diagnose unsupported -emit-llvm with fatbinary --- clang/include/clang/Basic/DiagnosticDriverKinds.td | 4 ++++ clang/lib/Driver/ToolChains/Cuda.cpp | 4 ++++ clang/test/Driver/cuda-emit-llvm-fatbinary-error.cu | 7 +++++++ 3 files changed, 15 insertions(+) create mode 100644 clang/test/Driver/cuda-emit-llvm-fatbinary-error.cu diff --git a/clang/include/clang/Basic/DiagnosticDriverKinds.td b/clang/include/clang/Basic/DiagnosticDriverKinds.td index db0f521b73544..8a7e70ddf118e 100644 --- a/clang/include/clang/Basic/DiagnosticDriverKinds.td +++ b/clang/include/clang/Basic/DiagnosticDriverKinds.td @@ -66,6 +66,10 @@ def err_drv_no_cuda_libdevice : Error< "via '--cuda-path', or pass '-nocudalib' to build without linking with " "libdevice">; +def err_drv_cuda_fatbinary_requires_elf_or_ptx : Error< + "cannot create CUDA fatbinary from LLVM IR/bitcode; use '--cuda-device-only' " + "or '--cuda-host-only', or compile device code to PTX/SASS (e.g. drop '-emit-llvm')">; + def err_drv_no_rocm_device_lib : Error< "cannot find ROCm device library%select{| for %1| for ABI version %1" "%select{|, which requires ROCm %3 or higher}2}0; provide its path via " diff --git a/clang/lib/Driver/ToolChains/Cuda.cpp b/clang/lib/Driver/ToolChains/Cuda.cpp index 6cc73ff5fc1f6..6b6e38def2ed7 100644 --- a/clang/lib/Driver/ToolChains/Cuda.cpp +++ b/clang/lib/Driver/ToolChains/Cuda.cpp @@ -546,6 +546,10 @@ void NVPTX::FatBinary::ConstructJob(Compilation &C, const JobAction &JA, CmdArgs.push_back("-g"); for (const auto &II : Inputs) { + if (II.getType() == types::TY_LLVM_BC || II.getType() == types::TY_LLVM_IR) { + C.getDriver().Diag(diag::err_drv_cuda_fatbinary_requires_elf_or_ptx); + return; + } auto *A = II.getAction(); assert(A->getInputs().size() == 1 && "Device offload action is expected to have a single input"); diff --git a/clang/test/Driver/cuda-emit-llvm-fatbinary-error.cu b/clang/test/Driver/cuda-emit-llvm-fatbinary-error.cu new file mode 100644 index 0000000000000..ee007e461f6e9 --- /dev/null +++ b/clang/test/Driver/cuda-emit-llvm-fatbinary-error.cu @@ -0,0 +1,7 @@ +// RUN: not %clangxx -### --target=x86_64-linux-gnu -S -emit-llvm --cuda-gpu-arch=sm_52 \ +// RUN: --cuda-path=%S/Inputs/CUDA/usr/local/cuda %s 2>&1 | FileCheck %s +// +// CHECK: error: cannot create CUDA fatbinary from LLVM IR/bitcode +// CHECK-NOT: "--create" + +__global__ void k() {} _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
