[libclc] Restore CLC_NATIVE_CPU define dropped by 6ef96710068e
Commit 6ef96710068e ("[libclc] Canonicalize 'clspv' to the
'spirv-unknown-vulkan' triple (#196351)") removed the generic
MACRO_ARCH / `string(TOUPPER "CLC_${MACRO_ARCH}" target_define)`
mechanism and the `${target_define}` compile definition, replacing it
with an explicit CLC_SPIRV define only. This silently dropped the
CLC_NATIVE_CPU define for the native_cpu target.
CLC_NATIVE_CPU is consumed in exactly one place,
libclc/clc/include/clc/clcfunc.h, where it gates the
__attribute__((libclc_call)) calling convention on the builtins:
#elif defined(CLC_NATIVE_CPU)
#define _CLC_DEF __attribute__((always_inline)) __attribute__((libclc_call))
#define _CLC_DECL __attribute__((libclc_call))
libclc_call maps to CC_SpirFunction (NativeCPUTargetInfo::
getLibclcCallingConv), which passes small vectors such as
vec<double,3> by value. Device code declares the matching __spirv_ocl_*
builtins with the same SpirFunction CC via SemaLookup.cpp. Without the
define, the native_cpu builtins were built with the default host ABI
(byval pointer) instead, so the library definition and the device-side
call sites disagreed on the ABI. Geometric/common builtins taking small
vectors (cross, dot, distance, normalize, ...) therefore received
garbage arguments and returned zeros.
Proof: compiling a single function for -triple native_cpu with
__attribute__((libclc_call)) emits `<3 x double> f(<3 x double>, ...)`
(by value); without it emits `f(ptr byval(<3 x double>), ...)`.
Fix: define CLC_NATIVE_CPU for the native_cpu arch branch in
libclc/CMakeLists.txt, mirroring the old target_define behavior. Only
clcfunc.h consumes this macro, so no other target is affected.
Symptom / regression vs origin/sycl on -fsycl-targets=native_cpu
(sycl::cross etc. returned {0,0,0}); the following E2E tests
assert-failed and now pass:
SYCL :: Basic/built-ins/const_vec_common.cpp
SYCL :: Basic/built-ins/const_vec_geometric.cpp
SYCL :: Basic/built-ins/marray_geometric.cpp
SYCL :: Basic/built-ins/vec_common.cpp
SYCL :: Basic/built-ins/vec_geometric.cpp
SYCL :: Basic/vector/int-convert.cpp
SYCL :: DeviceLib/built-ins/ext_native_math.cpp
SYCL :: DeviceLib/built-ins/vector_geometric.cpp
SYCL :: DeviceLib/built-ins/vector_relational.cpp
SYCL :: DeviceLib/math_test_marray_vec.cpp
SYCL :: Regression/mad_sat.cpp
Co-Authored-By: Claude Opus 4.8 (1M context) <noreply@anthropic.com>
