llvm
llvm copied to clipboard
`sycl::kernel_handler::get_specialization_constant` calls `__sycl_getComposite2020SpecConstantValue` on enumerated `value_type`
Describe the bug
sycl::kernel_handler::get_specialization_constant
calls __sycl_getComposite2020SpecConstantValue
instead of __sycl_getScalar2020SpecConstantValue
. Enumerated types have the same size, value representation and alignment requirements as their underlying integral types, so we could be simply using the scalar version of the builtin.
To Reproduce
Compile the following code:
#include <sycl/sycl.hpp>
enum class my_enum { a, b, c };
constexpr sycl::specialization_id<my_enum> val{my_enum::a};
SYCL_EXTERNAL my_enum f(sycl::kernel_handler kh) {
return kh.get_specialization_constant<val>();
}
clang++ -fsycl -fsycl-device-only -o - -S -emit-llvm -O3
You'll see the @_Z40__sycl_getComposite2020SpecConstantValueI7my_enumET_PKcPKvS5_
function being called to get the specialization constant value instead of the scalar version of the function.
Environment (please complete the following information):
- DPC++ version:
effbbabcd6e3
Additional context
This should have the exact same semantics, but using the scalar builtin leads to a simpler compilation path.
Changing the check in kernel_handler.hpp
so that the scalar version accepts enumerated types might be enough.
@victor-eds please, attach the code.
@victor-eds please, attach the code.
Added inline
Some massaging of our enable_if
s is required here:
https://github.com/intel/llvm/blob/5fae0aaa4f932c6ecddb052a0cb3d873ff2eee64/sycl/include/sycl/kernel_handler.hpp#L66-L85
We have a simple switch between compound
and fundamental
types, but we should probably use less C++-y terms like composite
and scalar
to treat more data types as scalar ones, even though they are compound
per C++
Some massaging of our
enable_if
s is required here:https://github.com/intel/llvm/blob/5fae0aaa4f932c6ecddb052a0cb3d873ff2eee64/sycl/include/sycl/kernel_handler.hpp#L66-L85
We have a simple switch between
compound
andfundamental
types, but we should probably use less C++-y terms likecomposite
andscalar
to treat more data types as scalar ones, even though they arecompound
per C++
According to cppreference, enums are scalar types. Maybe using std::is_scalar
would fix this?
Fixed by #12925