intel / intel/llvm

Improper handling of presumably invalid host and device target combinations

Open
#18,304 2 comments 0 reactions 2 assignees Claimed by @schittir View on GitHub
bug
Dominant language
LLVM
Stars
1.5k
Forks
854
Avg merge
3d 17h
Merged PRs (30d)
137

Description

### Describe the bug

Appropriate error handling for what are presumed to be invalid host and device target combinations seems to be missing. To ascertain this, I exercised the following host targets:
- i686-unknown-linux-gnu
- x86_64-unknown-linux-gnu
- i686-pc-windows-msvc
- x86_64-pc-windows-msvc

with each of the following device targets:
- amdgcn-amd-amdhsa
- nvptx-nvidia-cuda
- nvptx64-nvidia-cuda
- spir-unknown-unknown
- spir64-unknown-unknown
- spirv32-unknown-unknown
- spirv64-unknown-unknown

My test was limited to compiling the following source file (`t.cpp`):
```
#include
void f(sycl::handler &h) {
h.single_task([]{});
}
```

using a command line such as:
```
$ clang -c -fsycl -target -fsycl-targets= t.cpp
```

where `` is empty or:
- for amdgcn-amd-amdhsa: `-Xsycl-target-backend --offload-arch=gfx906 -fno-sycl-libspirv -nogpulib`
- for nvptx64-nvidia-cuda: `-fno-sycl-libspirv -nocudalib`

The use of `-fno-sycl-libspirv` with the amdgcn and nvptx64 targets was inspired by diagnostics like the following issued due to the lack of a CUDA or HIP SDK installation.
```
clang: error: cannot find 'remangled-l64-signed_char.libspirv-nvptx64-nvidia-cuda.bc'; provide path to libspirv library via '-fsycl-libspirv-path', or pass '-fno-sycl-libspirv' to build without linking with libspirv
```
However, the use of that option triggered a contradictory warning:
```
clang: warning: '-fno-sycl-libspirv' should not be used with target 'nvptx64-nvidia-cuda'; libspirv is required for correct behavior [-Wno-libspirv-hip-cuda]
```

Compiler invocations that target a Linux host were exercised on a 64-bit Linux system. Invocations that target a Windows host were exercised on a 64-bit Windows system.

Compilation was successful for the following host/device combinations:
- i686-unknown-linux-gnu:
- nvptx64-nvidia-cuda: This is surprising as I would not expect a 32-bit host to work with a 64-bit device.
- spir-unknown-unknown
- spirv32-unknown-unknown
- x86_64-unknown-linux-gnu:
- amdgcn-amd-amdhsa
- nvptx64-nvidia-cuda
- spir64-unknown-unknown
- spirv64-unknown-unknown
- i686-pc-windows-msvc:
- nvptx64-nvidia-cuda: This is surprising as I would not expect a 32-bit host to work with a 64-bit device.
- x86_64-pc-windows-msvc:
- amdgcn-amd-amdhsa
- nvptx64-nvidia-cuda
- spir64-unknown-unknown
- spirv64-unknown-unknown

Compilation failed with an appropriate diagnostic for the following combinations:
- *:
- nvptx-nvidia-cuda: This device target fails for all host targets; presumably because
it specifies a 32-bit device target and CUDA has dropped 32-bit support.
```
clang: error: SYCL target is invalid: 'nvptx-nvidia-cuda'
```

Compilation failed for other combinations with the behavior described below.
- i686-unknown-linux-gnu:
- amdgcn-amd-amdhsa:
- ```
clang-21: /.../llvm/clang/lib/AST/Expr.cpp:1053: clang::IntegerLiteral::IntegerLiteral(const ASTContext &, const llvm::APInt &, QualType, SourceLocation): Assertion `V.getBitWidth() == C.getIntWidth(type) && "Integer type is not the correct size for constant."' failed.`
```
- spir64-unknown-unknown:
- Various compilation errors followed by:
```
clang-21: /.../llvm/clang/lib/AST/Expr.cpp:1053: clang::IntegerLiteral::IntegerLiteral(const ASTContext &, const llvm::APInt &, QualType, SourceLocation): Assertion `V.getBitWidth() == C.getIntWidth(type) && "Integer type is not the correct size for constant."' failed.
```
- spirv64-unknown-unknown:
- Various compilation errors followed by:
```
clang-21: /.../llvm/clang/lib/AST/Expr.cpp:1053: clang::IntegerLiteral::IntegerLiteral(const ASTContext &, const llvm::APInt &, QualType, SourceLocation): Assertion `V.getBitWidth() == C.getIntWidth(type) && "Integer type is not the correct size for constant."' failed.
```
- x86_64-unknown-linux-gnu:
- spir-unknown-unknown:
- Various compilation errors due to `static_assert` failures in SYCL header files.
```
/.../bin/../include/sycl/detail/ur.hpp:136:17: error: static assertion failed due to requirement 'sizeof(_ze_context_handle_t *) == sizeof(unsigned long)': assert: cast failed size check
```
- spirv32-unknown-unknown:
- Various compilation errors due to `static_assert` failures in SYCL header files.
```
/.../bin/../include/sycl/detail/ur.hpp:136:17: error: static assertion failed due to requirement 'sizeof(_ze_context_handle_t *) == sizeof(unsigned long)': assert: cast failed size check
```
- i686-pc-windows-msvc:
- amdgcn-amd-amdhsa:
- Various compilation failures in Microsoft standard library header files.
- spir-unknown-unknown:
- Various compilation failures in Microsoft standard library header files.
- spir64-unknown-unknown:
- ```
Assertion failed: HT.getArch() == llvm::Triple::x86_64 && "Unsupported host architecture", file D:/.../clang/lib/Basic/Targets.cpp, line 754
```
- spirv32-unknown-unknown:
- Various compilation failures in Microsoft standard library header files.
- spirv64-unknown-unknown:
- ```
Assertion failed: HT.getArch() == llvm::Triple::x86_64 && "Unsupported host architecture", file D:/.../clang/lib/Basic/Targets.cpp, line 824
```
- x86_64-pc-windows-msvc:
- spir-unknown-unknown:
- ```
Assertion failed: HT.getArch() == llvm::Triple::x86 && "Unsupported host architecture", file D:/.../clang/lib/Basic/Targets.cpp, line 724
```
- spirv32-unknown-unknown:
- ```
Assertion failed: HT.getArch() == llvm::Triple::x86 && "Unsupported host architecture", file D:/.../clang/lib/Basic/Targets.cpp, line 803
```

This issue has been filed to request that invalid/unsupported host/device target combinations be appropriately detected and diagnosed by the Clang driver.

### To reproduce

See the issue description.

### Environment

See the issue description.

### Additional context

_No response_

Contributor guide

Open the contributing guide

Assessment

This issue has not been assessed yet.

Get new issues in your inbox

A short digest of beginner-friendly GitHub issues.