Merge pull request triSYCL#151 from airlied/cts-select_device

Implement select device

This series allows triSYCL to build the cts test_device_selector and execute it.

It unfortunately has to add a dummy `half` implementation and some basic bits of the program object, in order just to get the CTS common code to compile. Otherwise it fixes up the platform and device_selector code.

include/CL/sycl.hpp

@@ -46,6 +46,7 @@
 #include "CL/sycl/event.hpp"
 #include "CL/sycl/exception.hpp"
 #include "CL/sycl/group.hpp"
+#include "CL/sycl/half.hpp"
 #include "CL/sycl/handler.hpp"
 #include "CL/sycl/h_item.hpp"
 #include "CL/sycl/id.hpp"
@@ -59,6 +60,7 @@
 #include "CL/sycl/pipe.hpp"
 #include "CL/sycl/pipe_reservation.hpp"
 #include "CL/sycl/platform.hpp"
+#include "CL/sycl/program.hpp"
 #include "CL/sycl/queue.hpp"
 #include "CL/sycl/range.hpp"
 #include "CL/sycl/static_pipe.hpp"
@@ -67,6 +69,7 @@
 // Some includes at the end to break some dependencies
 #include "CL/sycl/device_selector/detail/device_selector_tail.hpp"
 #include "CL/sycl/device/detail/device_tail.hpp"
+#include "CL/sycl/platform/detail/platform_tail.hpp"
 #include "CL/sycl/platform/detail/host_platform_tail.hpp"
 #ifdef TRISYCL_OPENCL
 #include "CL/sycl/platform/detail/opencl_platform_tail.hpp"

include/CL/sycl/device_selector.hpp

@@ -9,9 +9,6 @@
     License. See LICENSE.TXT for details.
 */
 
-#include "CL/sycl/detail/unimplemented.hpp"
-#include "CL/sycl/device.hpp"
-
 namespace cl {
 namespace sycl {
 
@@ -25,17 +22,15 @@ namespace sycl {
 */
 class device_selector {
 
+  device select_device(vector_class<platform> platforms) const;
 public:
 
   /** Returns a selected device using the functor operator defined in
       sub-classes operator()(const device &dev)
 
       \todo Remove this from specification
   */
-  void /* device */ select_device() const {
-    //    return {};
-  }
-
+  device select_device() const;
 
   /**  This pure virtual operator allows the customization of device
        selection.

include/CL/sycl/device_selector/detail/device_selector_tail.hpp

@@ -23,7 +23,32 @@ namespace sycl {
 /** \addtogroup execution Platforms, contexts, devices and queues
     @{
 */
+inline device device_selector::select_device(vector_class<platform> platforms) const {
+  int best_id = -1;
+  int best_score = -1;
+  int i = 0;
+  device best_device;
+
+  for (auto &platform : platforms) {
+    for (auto &dev : platform.get_devices()) {
+      int score = operator()(dev);
+      if (score > best_score) {
+        best_id = 1;
+        best_device = dev;
+        best_score = score;
+      }
+      ++i;
+    }
+  }
+  if (best_id == -1)
+    throw cl::sycl::runtime_error("Could not find device");
 
+  return best_device;
+}
+
+inline device device_selector::select_device() const  {
+  return select_device(platform::get_platforms());
+}
 
 /** A device selector by device_type
 

include/CL/sycl/half.hpp

@@ -0,0 +1,24 @@
+#ifndef TRISYCL_SYCL_HALF_HPP
+#define TRISYCL_SYCL_HALF_HPP
+
+/** \file SYCL Half support - dummy
+    This file is distributed under the University of Illinois Open Source
+    License. See LICENSE.TXT for details.
+ */
+
+/* Dummy half implementation - this is enough to get the CTS to build. */
+namespace cl {
+namespace sycl {
+
+class half {
+public:
+  half(int) {};
+
+  bool operator>(const half &h1) { return false; };
+};
+
+}
+
+}
+
+#endif

include/CL/sycl/platform.hpp

@@ -68,7 +68,7 @@ class platform
 
       Retain a reference to the OpenCL platform.
   */
-  platform(cl_platform_id platform_id)
+  explicit platform(cl_platform_id platform_id)
     : platform { boost::compute::platform { platform_id } } {}
 
 
@@ -88,10 +88,7 @@ class platform
 
        Returns errors via the SYCL exception class.
   */
-  explicit platform(const device_selector &dev_selector) {
-    detail::unimplemented();
-  }
-
+  explicit platform(const device_selector &dev_selector);
 
 #ifdef TRISYCL_OPENCL
   /** Returns the cl_platform_id of the underlying OpenCL platform
@@ -177,10 +174,7 @@ class platform
       \return the device list
   */
   vector_class<device>
-  get_devices(info::device_type device_type = info::device_type::all) const {
-    return implementation->get_devices(device_type);
-  }
-
+  get_devices(info::device_type device_type = info::device_type::all) const;
 };
 
 /// @} to end the execution Doxygen group

include/CL/sycl/platform/detail/host_platform.hpp

@@ -114,7 +114,7 @@ auto static constexpr platform_extensions = "Xilinx_blocking_pipes";
       \return the device list
   */
   vector_class<cl::sycl::device>
-  get_devices(info::device_type device_type) const override;
+  get_devices(const device_selector &device_selector) const override;
 
 };
 

include/CL/sycl/platform/detail/host_platform_tail.hpp

@@ -28,11 +28,11 @@ namespace detail {
     \return the device list
 */
 vector_class<cl::sycl::device>
-inline host_platform::get_devices(info::device_type device_type) const {
+inline host_platform::get_devices(const device_selector &device_selector) const {
   /** If \c get_devices is called with the host platform
       and the right device type, returns the host_device.
   */
-  if (device_type_selector { device_type }(cl::sycl::device {}) > 0)
+  if (device_selector(cl::sycl::device {}) > 0)
     // Return 1 default device, i.e. the host device
     return { {} };
   else

include/CL/sycl/platform/detail/opencl_platform.hpp

@@ -97,7 +97,7 @@ class opencl_platform : public detail::platform {
       \return the device list
   */
   vector_class<cl::sycl::device>
-  get_devices(info::device_type device_type) const override;
+  get_devices(const device_selector &device_selector) const override;
 
 private:
 

include/CL/sycl/platform/detail/opencl_platform_tail.hpp

@@ -29,9 +29,8 @@ namespace detail {
     \return the device list
 */
 vector_class<cl::sycl::device>
-inline opencl_platform::get_devices(info::device_type device_type) const {
+inline opencl_platform::get_devices(const device_selector &device_selector) const {
   vector_class<cl::sycl::device> devices;
-    device_type_selector ds { device_type };
   // Add the desired OpenCL devices
   for (const auto &d : get_boost_compute().devices()) {
     // Get the SYCL device from the Boost Compute device
@@ -40,7 +39,7 @@ inline opencl_platform::get_devices(info::device_type device_type) const {
        By calling devices on the \c boost::compute::platform we know that
        we iterate only over the device belonging to the current platform,
     */
-    if (ds(sycl_dev) > 0)
+    if (device_selector(sycl_dev) > 0)
       devices.push_back(sycl_dev);
   }
 

include/CL/sycl/platform/detail/platform.hpp

@@ -17,6 +17,7 @@ namespace cl {
 namespace sycl {
 
 class device;
+class device_selector;
 
 namespace detail {
 
@@ -52,14 +53,12 @@ class platform {
 
   /** Get all the available devices for this platform
 
-      \param[in] device_type is the device type to filter the selection
-      or \c info::device_type::all by default to return all the
-      devices
+      \param[in] device_selector is used to filter the selection.
 
       \return the device list
   */
   virtual vector_class<device>
-  get_devices(info::device_type device_type) const = 0;
+  get_devices(const device_selector &device_selector) const = 0;
 
   // Virtual to call the real destructor
   virtual ~platform() {}

include/CL/sycl/platform/detail/platform_tail.hpp

@@ -0,0 +1,40 @@
+#ifndef TRISYCL_SYCL_PLATFORM_TAIL_HPP
+#define TRISYCL_SYCL_PLATFORM_TAIL_HPP
+
+/** \file The ending part of the OpenCL SYCL platform
+
+    This is here to break a dependence between platform and device_selector.
+    This file is distributed under the University of Illinois Open Source
+    License. See LICENSE.TXT for details.
+*/
+
+namespace cl {
+namespace sycl {
+
+inline
+platform::platform(const device_selector &device_selector) {
+  platform host_platform = {};
+#ifdef TRISYCL_OPENCL
+  if (host_platform.implementation->get_devices(device_selector).empty()) {
+    for (const auto &d : boost::compute::system::platforms()) {
+      auto clplatform = cl::sycl::platform { d };
+      auto devices = clplatform.implementation->get_devices(device_selector);
+      if (!devices.empty()) {
+        *this = std::move(clplatform);
+        return;
+      }
+    }
+  }
+#endif
+  *this = std::move(host_platform);
+}
+
+inline vector_class<device>
+platform::get_devices(info::device_type device_type) const {
+  return implementation->get_devices(device_type_selector { device_type });
+}
+
+}
+}
+
+#endif

include/CL/sycl/program.hpp

@@ -0,0 +1,103 @@
+#ifndef TRISYCL_SYCL_PROGRAM_HPP
+#define TRISYCL_SYCL_PROGRAM_HPP
+
+/** \file The OpenCL SYCL program
+
+    Dave Airlie
+
+    This file is distributed under the University of Illinois Open Source
+    License. See LICENSE.TXT for details.
+*/
+
+#ifdef TRISYCL_OPENCL
+#include <boost/compute.hpp>
+#endif
+
+#include "CL/sycl/detail/debug.hpp"
+#include "CL/sycl/detail/shared_ptr_implementation.hpp"
+#include "CL/sycl/detail/unimplemented.hpp"
+#include "CL/sycl/kernel.hpp"
+#include "CL/sycl/program/detail/program.hpp"
+
+namespace cl {
+namespace sycl {
+
+enum class program_state {
+  none,
+  compiled,
+  linked
+};
+
+/** SYCL program
+
+    \todo To be implemented
+
+*/
+
+class program
+    /* Use the underlying kernel implementation that can be shared in
+     the SYCL model */
+  : public detail::shared_ptr_implementation<program, detail::program> {
+
+  // The type encapsulating the implementation
+  using implementation_t = typename program::shared_ptr_implementation;
+
+  // The handler class uses the implementation
+  friend class handler;
+
+  // Allows the comparison operation to access the implementation
+  friend implementation_t;
+
+ public:
+
+  // Make the implementation member directly accessible in this class
+  using implementation_t::implementation;
+
+  program() = delete;
+
+#ifdef TRISYCL_OPENCL
+
+  program(const context &context, cl_program p) {}
+
+  cl_program get() const {
+    return implementation->get();
+  }
+
+  template <typename kernelT>
+  bool has_kernel() const;
+
+  bool has_kernel(string_class kernelName) const;
+
+  template <typename kernelT>
+  void compile_with_kernel_type(string_class compileOptions = "")
+  {
+
+  }
+
+  template <typename kernelT>
+  void build_with_kernel_type(string_class buildOptions = "")
+  {
+  }
+
+  template <typename kernelT>
+  cl::sycl::kernel get_kernel() const;
+
+  kernel get_kernel(string_class kernelName) const {
+    return NULL;
+  }
+#endif
+};
+
+}
+
+}
+
+/*
+    # Some Emacs stuff:
+    ### Local Variables:
+    ### ispell-local-dictionary: "american"
+    ### eval: (flyspell-prog-mode)
+    ### End:
+*/
+
+#endif