[SYCL] Update calls_indirectly property

sycl/include/sycl/ext/oneapi/experimental/virtual_functions.hpp

@@ -19,16 +19,16 @@ template <typename Set>
 inline constexpr indirectly_callable_key::value_t<Set> indirectly_callable_in;
 
 struct calls_indirectly_key {
-  template <typename First = void, typename... SetIds>
+  template <typename... SetIds>
   using value_t =
       sycl::ext::oneapi::experimental::property_value<calls_indirectly_key,
-                                                      First, SetIds...>;
+                                                      SetIds...>;
 };
 
 inline constexpr calls_indirectly_key::value_t<void> assume_indirect_calls;
 
-template <typename First, typename... Rest>
-inline constexpr calls_indirectly_key::value_t<First, Rest...>
+template <typename... SetIds>
+inline constexpr calls_indirectly_key::value_t<SetIds...>
     assume_indirect_calls_to;
 
 template <> struct is_property_key<indirectly_callable_key> : std::true_type {};
@@ -60,16 +60,54 @@ struct PropertyMetaInfo<indirectly_callable_key::value_t<Set>> {
 #endif
 };
 
-template <typename First, typename... Rest>
-struct PropertyMetaInfo<calls_indirectly_key::value_t<First, Rest...>> {
-  static_assert(
-      sizeof...(Rest) == 0,
-      "assume_indirect_calls_to property only supports a single set for now");
+// Helper to concatenate several lists of characters into a single string.
+// Lists are separated from each other with comma within the resulting string.
+template <typename List, typename... Rest> struct ConcatenateCharsToStr;
+
+// Specialization for a single list
+template <char... Chars> struct ConcatenateCharsToStr<CharList<Chars...>> {
+  static constexpr char value[] = {Chars..., '\0'};
+};
+
+// Specialization for two lists
+template <char... Chars, char... CharsToAppend>
+struct ConcatenateCharsToStr<CharList<Chars...>, CharList<CharsToAppend...>>
+    : ConcatenateCharsToStr<CharList<Chars..., ',', CharsToAppend...>> {};
+
+// Specialization for the case when there are more than two lists
+template <char... Chars, char... CharsToAppend, typename... Rest>
+struct ConcatenateCharsToStr<CharList<Chars...>, CharList<CharsToAppend...>,
+                             Rest...>
+    : ConcatenateCharsToStr<CharList<Chars..., ',', CharsToAppend...>,
+                            Rest...> {};
+
+// Helper to convert type T to a list of characters representing the type (its
+// mangled name).
+template <typename T, size_t... Indices> struct StableNameToCharsHelper {
+#ifdef __SYCL_DEVICE_ONLY__
+  using chars = CharList<__builtin_sycl_unique_stable_name(T)[Indices]...>;
+#else
+  using chars = CharList<>;
+#endif
+};
+
+// Wrapper helper for the struct above
+template <typename T, typename Sequence> struct StableNameToChars;
+
+// Specialization of that wrapper helper which accepts sequence of integers
+template <typename T, size_t... Indices>
+struct StableNameToChars<T, std::integer_sequence<size_t, Indices...>>
+    : StableNameToCharsHelper<T, Indices...> {};
+
+template <typename... SetIds>
+struct PropertyMetaInfo<calls_indirectly_key::value_t<SetIds...>> {
   static constexpr const char *name = "calls-indirectly";
   static constexpr const char *value =
 #ifdef __SYCL_DEVICE_ONLY__
-      // FIXME: we should handle Rest... here as well
-      __builtin_sycl_unique_stable_name(First);
+      ConcatenateCharsToStr<typename StableNameToChars<
+          SetIds,
+          std::make_index_sequence<__builtin_strlen(
+              __builtin_sycl_unique_stable_name(SetIds))>>::chars...>::value;
 #else
       "";
 #endif

sycl/test/virtual-functions/calls-indirectly-ir.cpp

@@ -11,16 +11,12 @@
 // CHECK: define {{.*}}KInt{{.*}} #[[#ATTR_SET_INT:]]
 // CHECK: define {{.*}}KVoid{{.*}} #[[#ATTR_SET_DEFAULT]]
 // CHECK: define {{.*}}KUserDefined{{.*}} #[[#ATTR_SET_USER_DEFINED:]]
-// TODO: update the check below
-// As of now calls_indirectly_property takes into account only the first
-// template argument ignoring the rest. This will be fixed in a follow-up
-// patches and the test should be updated to reflect that, because current
-// behavior is not correct.
-// CHECK-disabled: define {{.*}}KMultiple{{.*}} #[[#ATTR_SET_INT]]
+// CHECK: define {{.*}}KMultiple{{.*}} #[[#ATTR_SET_MULTIPLE:]]
 //
 // CHECK-DAG: attributes #[[#ATTR_SET_DEFAULT]] {{.*}} "calls-indirectly"="_ZTSv"
 // CHECK-DAG: attributes #[[#ATTR_SET_INT]] {{.*}} "calls-indirectly"="_ZTSi"
 // CHECK-DAG: attributes #[[#ATTR_SET_USER_DEFINED]] {{.*}} "calls-indirectly"="_ZTS12user_defined"
+// CHECK-DAG: attributes #[[#ATTR_SET_MULTIPLE]] {{.*}} "calls-indirectly"="_ZTSi,_ZTS12user_defined"
 
 #include <sycl/sycl.hpp>
 
@@ -45,17 +41,14 @@ int main() {
   oneapi::properties props_void{oneapi::assume_indirect_calls_to<void>};
   oneapi::properties props_user_defined{
       oneapi::assume_indirect_calls_to<user_defined>};
-  // assume_indirect_calls_to is currently limited to a single set, so this test
-  // is disabled.
-  // FIXME: re-enable once the restriction is lifted.
-  // oneapi::properties props_multiple{
-  //    oneapi::assume_indirect_calls_to<int, user_defined>};
+  oneapi::properties props_multiple{
+     oneapi::assume_indirect_calls_to<int, user_defined>};
 
   q.single_task<KEmpty>(props_empty, [=]() {});
   q.single_task<KInt>(props_int, [=]() {});
   q.single_task<KVoid>(props_void, [=]() {});
   q.single_task<KUserDefined>(props_user_defined, [=]() {});
-  // q.single_task<KMultiple>(props_multiple, [=]() {});
+  q.single_task<KMultiple>(props_multiple, [=]() {});
 
   return 0;
 }

sycl/test/virtual-functions/properties-positive.cpp

@@ -52,17 +52,14 @@ int main() {
   oneapi::properties props_void{oneapi::assume_indirect_calls_to<void>};
   oneapi::properties props_int{oneapi::assume_indirect_calls_to<int>};
   oneapi::properties props_base{oneapi::assume_indirect_calls_to<Base>};
-  // assume_indirect_calls_to is currently limited to a single set, so this test
-  // is disabled.
-  // FIXME: re-enable once the restriction is lifted.
-  // oneapi::properties props_multiple{
-  //    oneapi::assume_indirect_calls_to<int, Base>};
+  oneapi::properties props_multiple{
+     oneapi::assume_indirect_calls_to<int, Base>};
 
   q.single_task(props_empty, [=]() {});
   q.single_task(props_void, [=]() {});
   q.single_task(props_int, [=]() {});
   q.single_task(props_base, [=]() {});
-  // q.single_task(props_multiple, [=]() {});
+  q.single_task(props_multiple, [=]() {});
 
   return 0;
 }