[SYCL] Add new aspect ext_oneapi_virtual_functions (#15577)

Spec: #10540

Author: KornevNikita

llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td

@@ -85,6 +85,7 @@ def AspectExt_oneapi_virtual_mem : Aspect<"ext_oneapi_virtual_mem">;
 def AspectExt_oneapi_cuda_cluster_group : Aspect<"ext_oneapi_cuda_cluster_group">;
 def AspectExt_intel_fpga_task_sequence : Aspect<"ext_intel_fpga_task_sequence">;
 def AspectExt_oneapi_atomic16 : Aspect<"ext_oneapi_atomic16">;
+def AspectExt_oneapi_virtual_functions : Aspect<"ext_oneapi_virtual_functions">;
 // Deprecated aspects
 def AspectInt64_base_atomics : Aspect<"int64_base_atomics">;
 def AspectInt64_extended_atomics : Aspect<"int64_extended_atomics">;
@@ -148,7 +149,8 @@ def : TargetInfo<"__TestAspectList",
     AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph, AspectExt_oneapi_private_alloca, 
     AspectExt_oneapi_queue_profiling_tag, AspectExt_oneapi_virtual_mem, AspectExt_oneapi_cuda_cluster_group, 
     AspectExt_intel_fpga_task_sequence,
-    AspectExt_oneapi_atomic16],
+    AspectExt_oneapi_atomic16,
+    AspectExt_oneapi_virtual_functions],
     []>;
 // This definition serves the only purpose of testing whether the deprecated aspect list defined in here and in SYCL RT
 // match.

sycl/include/sycl/device_aspect_macros.hpp

@@ -395,6 +395,11 @@
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_atomic16__ 0
 #endif
 
+#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_virtual_functions__
+//__SYCL_ASPECT(ext_oneapi_virtual_functions, 81)
+#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_virtual_functions__ 0
+#endif
+
 #ifndef __SYCL_ANY_DEVICE_HAS_host__
 // __SYCL_ASPECT(host, 0)
 #define __SYCL_ANY_DEVICE_HAS_host__ 0
@@ -779,3 +784,8 @@
 //__SYCL_ASPECT(ext_oneapi_oneapi_atomic16, 80)
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_atomic16__ 0
 #endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_virtual_functions__
+//__SYCL_ASPECT(ext_oneapi_virtual_functions, 81)
+#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_virtual_functions__ 0
+#endif

sycl/include/sycl/info/aspects.def

@@ -71,3 +71,4 @@ __SYCL_ASPECT(ext_oneapi_unique_addressing_per_dim, 77)
 __SYCL_ASPECT(ext_oneapi_bindless_images_sample_1d_usm, 78)
 __SYCL_ASPECT(ext_oneapi_bindless_images_sample_2d_usm, 79)
 __SYCL_ASPECT(ext_oneapi_atomic16, 80)
+__SYCL_ASPECT(ext_oneapi_virtual_functions, 81)

sycl/source/detail/device_impl.cpp

@@ -768,6 +768,13 @@ bool device_impl::has(aspect Aspect) const {
     // Likely L0 doesn't check it properly. Need to double-check.
     return has_extension("cl_ext_float_atomics");
   }
+  case aspect::ext_oneapi_virtual_functions: {
+    // TODO: move to UR like e.g. aspect::ext_oneapi_virtual_mem
+    backend BE = getBackend();
+    bool isCompatibleBE = BE == sycl::backend::ext_oneapi_level_zero ||
+                          BE == sycl::backend::opencl;
+    return (is_cpu() || is_gpu()) && isCompatibleBE;
+  }
   }
 
   return false; // This device aspect has not been implemented yet.

sycl/test-e2e/Basic/aspects.cpp

@@ -90,6 +90,9 @@ int main() {
     if (plt.has(aspect::ext_oneapi_atomic16)) {
       std::cout << "  ext_oneapi_atomic16" << std::endl;
     }
+    if (plt.has(aspect::ext_oneapi_virtual_functions)) {
+      std::cout << "  ext_oneapi_virtual_functions" << std::endl;
+    }
   }
   std::cout << "Passed." << std::endl;
   return 0;

sycl/test-e2e/VirtualFunctions/2/1/1/missing-overrides.cpp

@@ -1,6 +1,3 @@
-// UNSUPPORTED: cuda, hip, acc
-// FIXME: replace unsupported with an aspect check once we have it
-//
 // RUN: %{build} -o %t.out %helper-includes
 // RUN: %{run} %t.out
 

sycl/test-e2e/VirtualFunctions/2/1/1/more-complex-hierarchy.cpp

@@ -1,6 +1,3 @@
-// UNSUPPORTED: cuda, hip, acc
-// FIXME: replace unsupported with an aspect check once we have it
-//
 // RUN: %{build} -o %t.out %helper-includes
 // RUN: %{run} %t.out
 

sycl/test-e2e/VirtualFunctions/2/1/1/simple-hierarchy.cpp

@@ -1,6 +1,3 @@
-// UNSUPPORTED: cuda, hip, acc
-// FIXME: replace unsupported with an aspect check once we have it
-//
 // RUN: %{build} -o %t.out %helper-includes
 // RUN: %{run} %t.out
 

sycl/test-e2e/VirtualFunctions/2/2/single-construct-single-use.cpp

@@ -1,6 +1,3 @@
-// UNSUPPORTED: cuda, hip, acc
-// FIXME: replace unsupported with an aspect check once we have it
-//
 // RUN: %{build} -o %t.out %helper-includes
 // RUN: %{run} %t.out
 

sycl/test-e2e/VirtualFunctions/lit.local.cfg

@@ -4,3 +4,4 @@ import os
 # paths like "../../../helper.hpp" in them, so let's just register a
 # substitution to add directory with helper headers into include search path
 config.substitutions.append(("%helper-includes", "-I {}".format(os.path.dirname(os.path.abspath(__file__)))))
+config.required_features += ['aspect-ext_oneapi_virtual_functions']