program.hpp

Go to the documentation of this file.

//==--------------- program.hpp --- SYCL program ---------------------------==//
//
// 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
//
//===----------------------------------------------------------------------===//
 
#pragma once
 
#include <CL/sycl/context.hpp>
#include <CL/sycl/detail/export.hpp>
#include <CL/sycl/detail/kernel_desc.hpp>
#include <CL/sycl/detail/os_util.hpp>
#include <CL/sycl/info/info_desc.hpp>
#include <CL/sycl/kernel.hpp>
#include <CL/sycl/property_list.hpp>
#include <CL/sycl/stl.hpp>
#include <sycl/ext/oneapi/experimental/spec_constant.hpp>
 
__SYCL_INLINE_NAMESPACE(cl) {
namespace sycl {
 
// Forward declarations
class context;
class device;
namespace detail {
class program_impl;
}
 
enum class program_state { none = 0, compiled = 1, linked = 2 };
 
class __SYCL_EXPORT __SYCL2020_DEPRECATED(
    "program class is deprecated, use kernel_bundle instead") program {
public:
  program() = delete;
 
  explicit program(const context &Context, const property_list &PropList = {});
 
  program(const context &Context, std::vector<device> DeviceList,
          const property_list &PropList = {});
 
  program(std::vector<program> ProgramList, const property_list &PropList = {});
 
  program(std::vector<program> ProgramList, std::string LinkOptions,
          const property_list &PropList = {});
 
  program(const context &Context, cl_program ClProgram);
 
  program(const program &rhs) = default;
 
  program(program &&rhs) = default;
 
  program &operator=(const program &rhs) = default;
 
  program &operator=(program &&rhs) = default;
 
  bool operator==(const program &rhs) const { return impl == rhs.impl; }
 
  bool operator!=(const program &rhs) const { return impl != rhs.impl; }
 
  template <typename propertyT> bool has_property() const;
 
  template <typename propertyT> propertyT get_property() const;
 
  cl_program get() const;
 
  bool is_host() const;
 
  template <typename KernelT>
  void compile_with_kernel_type(std::string CompileOptions = "") {
    detail::OSModuleHandle M = detail::OSUtil::getOSModuleHandle(
        detail::KernelInfo<KernelT>::getName());
    compile_with_kernel_name(detail::KernelInfo<KernelT>::getName(),
                             CompileOptions, M);
  }
 
  void compile_with_source(std::string KernelSource,
                           std::string CompileOptions = "");
 
  template <typename KernelT>
  void build_with_kernel_type(std::string BuildOptions = "") {
    detail::OSModuleHandle M = detail::OSUtil::getOSModuleHandle(
        detail::KernelInfo<KernelT>::getName());
    build_with_kernel_name(detail::KernelInfo<KernelT>::getName(), BuildOptions,
                           M);
  }
 
  void build_with_source(std::string KernelSource,
                         std::string BuildOptions = "");
 
  void link(std::string LinkOptions = "");
 
  template <typename KernelT> bool has_kernel() const {
    return has_kernel(detail::KernelInfo<KernelT>::getName(),
                      /*IsCreatedFromSource*/ false);
  }
 
  bool has_kernel(std::string KernelName) const;
 
  template <typename KernelT> kernel get_kernel() const {
    return get_kernel(detail::KernelInfo<KernelT>::getName(),
                      /*IsCreatedFromSource*/ false);
  }
 
  kernel get_kernel(std::string KernelName) const;
 
  template <info::program param>
  typename info::param_traits<info::program, param>::return_type
  get_info() const;
 
  std::vector<std::vector<char>> get_binaries() const;
 
  context get_context() const;
 
  std::vector<device> get_devices() const;
 
  std::string get_compile_options() const;
 
  std::string get_link_options() const;
 
  std::string get_build_options() const;
 
  program_state get_state() const;
 
  template <typename ID, typename T>
  ext::oneapi::experimental::spec_constant<T, ID> set_spec_constant(T Cst) {
    constexpr const char *Name = detail::SpecConstantInfo<ID>::getName();
    static_assert(std::is_arithmetic<T>::value ||
                      (std::is_class<T>::value && std::is_pod<T>::value),
                  "unsupported specialization constant type");
#ifdef __SYCL_DEVICE_ONLY__
    (void)Cst;
    (void)Name;
    return ext::oneapi::experimental::spec_constant<T, ID>();
#else
    set_spec_constant_impl(Name, &Cst, sizeof(T));
    return ext::oneapi::experimental::spec_constant<T, ID>(Cst);
#endif // __SYCL_DEVICE_ONLY__
  }
 
  backend get_backend() const noexcept;
 
  template <backend BackendName>
  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
  auto get_native() const -> typename interop<BackendName, program>::type {
    return reinterpret_cast<typename interop<BackendName, program>::type>(
        getNative());
  }
 
private:
  pi_native_handle getNative() const;
  program(std::shared_ptr<detail::program_impl> impl);
 
  kernel get_kernel(std::string KernelName, bool IsCreatedFromSource) const;
 
  bool has_kernel(std::string KernelName, bool IsCreatedFromSource) const;
 
  void compile_with_kernel_name(std::string KernelName,
                                std::string CompileOptions,
                                detail::OSModuleHandle M);
 
  void build_with_kernel_name(std::string KernelName, std::string buildOptions,
                              detail::OSModuleHandle M);
 
  void set_spec_constant_impl(const char *Name, void *Data, size_t Size);
 
  std::shared_ptr<detail::program_impl> impl;
 
  template <class Obj>
  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
  template <class T>
  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
};
} // namespace sycl
} // __SYCL_INLINE_NAMESPACE(cl)
 
namespace std {
template <> struct hash<cl::sycl::program> {
  size_t operator()(const cl::sycl::program &prg) const {
    return hash<std::shared_ptr<cl::sycl::detail::program_impl>>()(
        cl::sycl::detail::getSyclObjImpl(prg));
  }
};
} // namespace std