Skip to content

[SYCL] Limit decorations to init function for annotated_arg/ptr classes #11881

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 6 commits into from
Nov 16, 2023
Merged

[SYCL] Limit decorations to init function for annotated_arg/ptr classes #11881

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
6 commits
Select commit Hold shift + click to select a range
6dc7111
move ptr decoration to init func args only
tiwaria1 Nov 13, 2023
34aa107
fix formatting
tiwaria1 Nov 14, 2023
92c71d7
run clang-format
tiwaria1 Nov 14, 2023
266b40f
clang-format again
tiwaria1 Nov 14, 2023
9dd8a13
add typename keyword
tiwaria1 Nov 14, 2023
dc73e96
clang-format
tiwaria1 Nov 14, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
16 changes: 5 additions & 11 deletions sycl/include/sycl/ext/oneapi/experimental/annotated_arg/annotated_arg.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -65,20 +65,15 @@ class __SYCL_SPECIAL_CLASS
__SYCL_TYPE(annotated_arg) annotated_arg<T *, detail::properties_t<Props...>> {
using property_list_t = detail::properties_t<Props...>;

#ifdef __SYCL_DEVICE_ONLY__
using global_pointer_t = typename decorated_global_ptr<T>::pointer;
#else
using global_pointer_t = T *;
#endif

global_pointer_t obj;
T *obj;

template <typename T2, typename PropertyListT> friend class annotated_arg;

#ifdef __SYCL_DEVICE_ONLY__
void __init([[__sycl_detail__::add_ir_attributes_kernel_parameter(
detail::PropertyMetaInfo<Props>::name...,
detail::PropertyMetaInfo<Props>::value...)]] global_pointer_t _obj) {
detail::PropertyMetaInfo<Props>::value...)]]
typename decorated_global_ptr<T>::pointer _obj) {
obj = _obj;
}
#endif
Expand Down Expand Up @@ -110,7 +105,7 @@ __SYCL_TYPE(annotated_arg) annotated_arg<T *, detail::properties_t<Props...>> {

annotated_arg(T *_ptr,
const property_list_t &PropList = properties{}) noexcept
: obj(global_pointer_t(_ptr)) {
: obj(_ptr) {
(void)PropList;
}

Expand All @@ -119,8 +114,7 @@ __SYCL_TYPE(annotated_arg) annotated_arg<T *, detail::properties_t<Props...>> {
// variadic properties. The same property in `Props...` and
// `PropertyValueTs...` must have the same property value.
template <typename... PropertyValueTs>
annotated_arg(T *_ptr, const PropertyValueTs &...props) noexcept
: obj(global_pointer_t(_ptr)) {
annotated_arg(T *_ptr, const PropertyValueTs &...props) noexcept : obj(_ptr) {
static constexpr bool has_same_properties = std::is_same<
property_list_t,
detail::merged_properties_t<property_list_t,
Expand Down
10 changes: 3 additions & 7 deletions sycl/include/sycl/ext/oneapi/experimental/annotated_ptr/annotated_ptr.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -190,7 +190,6 @@ __SYCL_TYPE(annotated_ptr) annotated_ptr<T, detail::properties_t<Props...>> {
using reference = sycl::ext::oneapi::experimental::annotated_ref<
T, typename unpack<filtered_properties>::type>;

#ifdef __SYCL_DEVICE_ONLY__
#ifdef __ENABLE_USM_ADDR_SPACE__
using global_pointer_t = std::conditional_t<
detail::IsUsmKindDevice<property_list_t>::value,
Expand All @@ -202,11 +201,8 @@ __SYCL_TYPE(annotated_ptr) annotated_ptr<T, detail::properties_t<Props...>> {
#else
using global_pointer_t = typename decorated_global_ptr<T>::pointer;
#endif // __ENABLE_USM_ADDR_SPACE__
#else
using global_pointer_t = T *;
#endif // __SYCL_DEVICE_ONLY__

global_pointer_t m_Ptr;
T *m_Ptr;

template <typename T2, typename PropertyListT> friend class annotated_ptr;

Expand Down Expand Up @@ -245,15 +241,15 @@ __SYCL_TYPE(annotated_ptr) annotated_ptr<T, detail::properties_t<Props...>> {

explicit annotated_ptr(T *Ptr,
const property_list_t & = properties{}) noexcept
: m_Ptr(global_pointer_t(Ptr)) {}
: m_Ptr(Ptr) {}

// Constructs an annotated_ptr object from a raw pointer and variadic
// properties. The new property set contains all properties of the input
// variadic properties. The same property in `Props...` and
// `PropertyValueTs...` must have the same property value.
template <typename... PropertyValueTs>
explicit annotated_ptr(T *Ptr, const PropertyValueTs &...props) noexcept
: m_Ptr(global_pointer_t(Ptr)) {
: m_Ptr(Ptr) {
static constexpr bool has_same_properties = std::is_same<
property_list_t,
detail::merged_properties_t<property_list_t,
Expand Down