[SYCL] Use public interfaces in function_pointer.hpp (#1097)

This patch is a part of effort to decouple SYCL Runtime library
interface from its actual implementation. The goal is to improve
SYCL ABI/API compatibility between different versions of library.

Two helper functions were introduced to eliminate calls to _impl
classes. This caused 3 tests fail. 2 tests were found to use
internal APIs. Missing include clauses were added to workaround
those fails. A more robust solution will be provided in separate
patches.

Signed-off-by: Alexander Batashev <alexander.batashev@intel.com>

sycl/include/CL/sycl/device.hpp

@@ -22,7 +22,6 @@ class device_selector;
 namespace detail {
 class device_impl;
 }
-
 class device {
 public:
   /// Constructs a SYCL device instance as a host device.

sycl/include/CL/sycl/intel/function_pointer.hpp

@@ -8,7 +8,6 @@
 
 #pragma once
 
-#include <CL/sycl/detail/program_impl.hpp>
 #include <CL/sycl/device.hpp>
 #include <CL/sycl/program.hpp>
 #include <CL/sycl/stl.hpp>
@@ -17,6 +16,10 @@
 
 __SYCL_INLINE namespace cl {
 namespace sycl {
+namespace detail {
+cl_ulong getDeviceFunctionPointerImpl(device &D, program &P,
+                                      const char *FuncName);
+}
 namespace intel {
 
 // This is a preview extension implementation, intended to provide early access
@@ -27,7 +30,7 @@ namespace intel {
 // products. If you are interested in using this feature in your software
 // product, please let us know!
 
-using device_func_ptr_holder_t = cl::sycl::cl_ulong;
+using device_func_ptr_holder_t = cl_ulong;
 
 /// \brief this function performs a cast from device_func_ptr_holder_t type
 /// to the provided function pointer type.
@@ -78,15 +81,7 @@ device_func_ptr_holder_t get_device_func_ptr(FuncType F, const char *FuncName,
         "Program must be built before passing to get_device_func_ptr");
   }
 
-  device_func_ptr_holder_t FPtr = 0;
-  // FIXME: return value must be checked here, but since we cannot yet check
-  // if corresponding extension is supported, let's silently ignore it here.
-  PI_CALL(piextGetDeviceFunctionPointer)(
-      detail::pi::cast<pi_device>(detail::getSyclObjImpl(D)->getHandleRef()),
-      detail::pi::cast<pi_program>(detail::getSyclObjImpl(P)->getHandleRef()),
-      FuncName, &FPtr);
-
-  return FPtr;
+  return detail::getDeviceFunctionPointerImpl(D, P, FuncName);
 }
 
 } // namespace intel

sycl/source/CMakeLists.txt

@@ -88,6 +88,7 @@ set(SYCL_SOURCES
     "event.cpp"
     "exception.cpp"
     "exception_list.cpp"
+    "function_pointer.cpp"
     "half_type.cpp"
     "handler.cpp"
     "kernel.cpp"

sycl/source/function_pointer.cpp

@@ -0,0 +1,30 @@
+//==----------- function_pointer.cpp --- SYCL Function pointers ------------==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <CL/sycl/intel/function_pointer.hpp>
+#include <CL/sycl/detail/device_impl.hpp>
+#include <CL/sycl/detail/program_impl.hpp>
+
+__SYCL_INLINE namespace cl {
+namespace sycl {
+namespace detail {
+intel::device_func_ptr_holder_t
+getDeviceFunctionPointerImpl(device &D, program &P, const char *FuncName) {
+  intel::device_func_ptr_holder_t FPtr = 0;
+  // FIXME: return value must be checked here, but since we cannot yet check
+  // if corresponding extension is supported, let's silently ignore it here.
+  PI_CALL(piextGetDeviceFunctionPointer)(
+      detail::pi::cast<pi_device>(detail::getSyclObjImpl(D)->getHandleRef()),
+      detail::pi::cast<pi_program>(detail::getSyclObjImpl(P)->getHandleRef()),
+      FuncName, &FPtr);
+  return FPtr;
+}
+
+} // namespace detail
+} // namespace sycl
+} // namespace cl

sycl/test/basic_tests/image_api.cpp

@@ -7,6 +7,9 @@
 // RUN: %ACC_RUN_PLACEHOLDER %t1.out
 
 #include <CL/sycl.hpp>
+// FIXME do not use internal methods in tests.
+#include <CL/sycl/detail/cg.hpp>
+#include <CL/sycl/detail/kernel_impl.hpp>
 
 #include <algorithm>
 #include <array>

sycl/test/kernel-and-program/cache.cpp

@@ -9,6 +9,8 @@
 //===----------------------------------------------------------------------===//
 
 #include <CL/sycl.hpp>
+// FIXME do not use internal methods in tests.
+#include <CL/sycl/detail/program_impl.hpp>
 
 namespace RT = cl::sycl::RT;
 namespace detail = cl::sycl::detail;

sycl/test/sub_group/common_ocl.cpp

@@ -16,7 +16,9 @@
 #include "helper.hpp"
 #include <CL/sycl.hpp>
 #include <cstring>
+#include <fstream>
 #include <iostream>
+
 using namespace cl::sycl;
 struct Data {
   unsigned int local_id;