Compute Library
 20.02.1
CLKernelLibrary.h
Go to the documentation of this file.
1 /*
2  * Copyright (c) 2016-2019 ARM Limited.
3  *
4  * SPDX-License-Identifier: MIT
5  *
6  * Permission is hereby granted, free of charge, to any person obtaining a copy
7  * of this software and associated documentation files (the "Software"), to
8  * deal in the Software without restriction, including without limitation the
9  * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
10  * sell copies of the Software, and to permit persons to whom the Software is
11  * furnished to do so, subject to the following conditions:
12  *
13  * The above copyright notice and this permission notice shall be included in all
14  * copies or substantial portions of the Software.
15  *
16  * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
17  * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
18  * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
19  * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
20  * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
21  * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
22  * SOFTWARE.
23  */
24 #ifndef ARM_COMPUTE_CLKERNELLIBRARY_H
25 #define ARM_COMPUTE_CLKERNELLIBRARY_H
26 
28 
29 #include <map>
30 #include <set>
31 #include <string>
32 #include <utility>
33 
34 namespace arm_compute
35 {
36 /** Build options */
37 class CLBuildOptions final
38 {
39  using StringSet = std::set<std::string>;
40 
41 public:
42  /** Default constructor. */
44  /** Adds option to the existing build option list
45  *
46  * @param[in] option Option to add
47  */
48  void add_option(std::string option);
49  /** Adds option if a given condition is true;
50  *
51  * @param[in] cond Condition to check
52  * @param[in] option Option to add if condition is true
53  */
54  void add_option_if(bool cond, std::string option);
55  /** Adds first option if condition is true else the second one
56  *
57  * @param[in] cond Condition to check
58  * @param[in] option_true Option to add if condition is true
59  * @param[in] option_false Option to add if condition is false
60  */
61  void add_option_if_else(bool cond, std::string option_true, std::string option_false);
62  /** Appends given build options to the current's objects options.
63  *
64  * @param[in] options Build options to append
65  */
66  void add_options(const StringSet &options);
67  /** Appends given build options to the current's objects options if a given condition is true.
68  *
69  * @param[in] cond Condition to check
70  * @param[in] options Option to add if condition is true
71  */
72  void add_options_if(bool cond, const StringSet &options);
73  /** Gets the current options list set
74  *
75  * @return Build options set
76  */
77  const StringSet &options() const;
78 
79 private:
80  StringSet _build_opts; /**< Build options set */
81 };
82 /** Program class */
83 class Program final
84 {
85 public:
86  /** Default constructor. */
87  Program();
88  /** Construct program from source file.
89  *
90  * @param[in] context CL context used to create the program.
91  * @param[in] name Program name.
92  * @param[in] source Program source.
93  */
94  Program(cl::Context context, std::string name, std::string source);
95  /** Construct program from binary file.
96  *
97  * @param[in] context CL context used to create the program.
98  * @param[in] device CL device for which the programs are created.
99  * @param[in] name Program name.
100  * @param[in] binary Program binary.
101  */
102  Program(cl::Context context, cl::Device device, std::string name, std::vector<unsigned char> binary);
103  /** Default Copy Constructor. */
104  Program(const Program &) = default;
105  /** Default Move Constructor. */
106  Program(Program &&) = default;
107  /** Default copy assignment operator */
108  Program &operator=(const Program &) = default;
109  /** Default move assignment operator */
110  Program &operator=(Program &&) = default;
111  /** Returns program name.
112  *
113  * @return Program's name.
114  */
115  std::string name() const
116  {
117  return _name;
118  }
119  /** User-defined conversion to the underlying CL program.
120  *
121  * @return The CL program object.
122  */
123  explicit operator cl::Program() const;
124  /** Build the given CL program.
125  *
126  * @param[in] program The CL program to build.
127  * @param[in] build_options Options to build the CL program.
128  *
129  * @return True if the CL program builds successfully.
130  */
131  static bool build(const cl::Program &program, const std::string &build_options = "");
132  /** Build the underlying CL program.
133  *
134  * @param[in] build_options Options used to build the CL program.
135  *
136  * @return A reference to itself.
137  */
138  cl::Program build(const std::string &build_options = "") const;
139 
140 private:
141  cl::Context _context; /**< Underlying CL context. */
142  cl::Device _device; /**< CL device for which the programs are created. */
143  bool _is_binary; /**< Create program from binary? */
144  std::string _name; /**< Program name. */
145  std::string _source; /**< Source code for the program. */
146  std::vector<unsigned char> _binary; /**< Binary from which to create the program. */
147 };
148 
149 /** Kernel class */
150 class Kernel final
151 {
152 public:
153  /** Default Constructor. */
154  Kernel();
155  /** Default Copy Constructor. */
156  Kernel(const Kernel &) = default;
157  /** Default Move Constructor. */
158  Kernel(Kernel &&) = default;
159  /** Default copy assignment operator */
160  Kernel &operator=(const Kernel &) = default;
161  /** Default move assignment operator */
162  Kernel &operator=(Kernel &&) = default;
163  /** Constructor.
164  *
165  * @param[in] name Kernel name.
166  * @param[in] program Built program.
167  */
168  Kernel(std::string name, const cl::Program &program);
169  /** Returns kernel name.
170  *
171  * @return Kernel's name.
172  */
173  std::string name() const
174  {
175  return _name;
176  }
177  /** Returns OpenCL kernel.
178  *
179  * @return OpenCL Kernel.
180  */
181  explicit operator cl::Kernel() const
182  {
183  return _kernel;
184  }
185 
186 private:
187  std::string _name; /**< Kernel name */
188  cl::Kernel _kernel; /**< OpenCL Kernel */
189 };
190 
191 /** CLKernelLibrary class */
192 class CLKernelLibrary final
193 {
194  using StringSet = std::set<std::string>;
195 
196 public:
197  /** Default Constructor. */
198  CLKernelLibrary();
199  /** Prevent instances of this class from being copied */
200  CLKernelLibrary(const CLKernelLibrary &) = delete;
201  /** Prevent instances of this class from being copied */
202  const CLKernelLibrary &operator=(const CLKernelLibrary &) = delete;
203  /** Access the KernelLibrary singleton.
204  * This method has been deprecated and will be removed in the next release.
205  * @return The KernelLibrary instance.
206  */
207  static CLKernelLibrary &get();
208  /** Initialises the kernel library.
209  *
210  * @param[in] kernel_path Path of the directory from which kernel sources are loaded.
211  * @param[in] context CL context used to create programs.
212  * @param[in] device CL device for which the programs are created.
213  */
214  void init(std::string kernel_path, cl::Context context, cl::Device device);
215  /** Sets the path that the kernels reside in.
216  *
217  * @param[in] kernel_path Path of the kernel.
218  */
219  void set_kernel_path(const std::string &kernel_path);
220  /** Gets the path that the kernels reside in.
221  */
222  std::string get_kernel_path();
223  /** Gets the source of the selected program.
224  *
225  * @param[in] program_name Program name.
226  *
227  * @return Source of the selected program.
228  */
229  std::string get_program_source(const std::string &program_name);
230  /** Sets the CL context used to create programs.
231  *
232  * @note Setting the context also resets the device to the
233  * first one available in the new context.
234  *
235  * @param[in] context A CL context.
236  */
237  void set_context(cl::Context context);
238 
239  /** Accessor for the associated CL context.
240  *
241  * @return A CL context.
242  */
243  cl::Context &context();
244 
245  /** Gets the CL device for which the programs are created. */
246  cl::Device &get_device();
247 
248  /** Sets the CL device for which the programs are created.
249  *
250  * @param[in] device A CL device.
251  */
252  void set_device(cl::Device device);
253 
254  /** Return the device version
255  *
256  * @return The content of CL_DEVICE_VERSION
257  */
258  std::string get_device_version();
259  /** Return the maximum number of compute units in the device
260  *
261  * @return The content of CL_DEVICE_MAX_COMPUTE_UNITS
262  */
263  cl_uint get_num_compute_units();
264  /** Creates a kernel from the kernel library.
265  *
266  * @param[in] kernel_name Kernel name.
267  * @param[in] build_options_set Kernel build options as a set.
268  *
269  * @return The created kernel.
270  */
271  Kernel create_kernel(const std::string &kernel_name, const StringSet &build_options_set = {}) const;
272  /** Find the maximum number of local work items in a workgroup can be supported for the kernel.
273  *
274  */
275  size_t max_local_workgroup_size(const cl::Kernel &kernel) const;
276  /** Return the default NDRange for the device.
277  *
278  */
279  cl::NDRange default_ndrange() const;
280 
281  /** Clear the library's cache of binary programs
282  */
283  void clear_programs_cache();
284 
285  /** Access the cache of built OpenCL programs */
286  const std::map<std::string, cl::Program> &get_built_programs() const;
287 
288  /** Add a new built program to the cache
289  *
290  * @param[in] built_program_name Name of the program
291  * @param[in] program Built program to add to the cache
292  */
293  void add_built_program(const std::string &built_program_name, const cl::Program &program);
294 
295  /** Returns true if FP16 is supported by the CL device
296  *
297  * @return true if the CL device supports FP16
298  */
299  bool fp16_supported() const;
300 
301  /** Returns true if int64_base_atomics extension is supported by the CL device
302  *
303  * @return true if the CL device supports int64_base_atomics extension
304  */
305  bool int64_base_atomics_supported() const;
306 
307 private:
308  /** Load program and its dependencies.
309  *
310  * @param[in] program_name Name of the program to load.
311  */
312  const Program &load_program(const std::string &program_name) const;
313  /** Concatenates contents of a set into a single string.
314  *
315  * @param[in] s Input set to concatenate.
316  *
317  * @return Concatenated string.
318  */
319  std::string stringify_set(const StringSet &s) const;
320 
321  cl::Context _context; /**< Underlying CL context. */
322  cl::Device _device; /**< Underlying CL device. */
323  std::string _kernel_path; /**< Path to the kernels folder. */
324  mutable std::map<std::string, const Program> _programs_map; /**< Map with all already loaded program data. */
325  mutable std::map<std::string, cl::Program> _built_programs_map; /**< Map with all already built program data. */
326  static const std::map<std::string, std::string> _kernel_program_map; /**< Map that associates kernel names with programs. */
327  static const std::map<std::string, std::string> _program_source_map; /**< Contains sources for all programs.
328  Used for compile-time kernel inclusion. >*/
329 };
330 } // namespace arm_compute
331 #endif /* ARM_COMPUTE_CLKERNELLIBRARY_H */
void set_kernel_path(const std::string &kernel_path)
Sets the path that the kernels reside in.
cl::Device & get_device()
Gets the CL device for which the programs are created.
Kernel()
Default Constructor.
void add_built_program(const std::string &built_program_name, const cl::Program &program)
Add a new built program to the cache.
std::string get_device_version()
Return the device version.
const StringSet & options() const
Gets the current options list set.
static CLKernelLibrary & get()
Access the KernelLibrary singleton.
CLBuildOptions()
Default constructor.
void set_context(cl::Context context)
Sets the CL context used to create programs.
Copyright (c) 2017-2020 ARM Limited.
cl_uint get_num_compute_units()
Return the maximum number of compute units in the device.
size_t max_local_workgroup_size(const cl::Kernel &kernel) const
Find the maximum number of local work items in a workgroup can be supported for the kernel.
const CLKernelLibrary & operator=(const CLKernelLibrary &)=delete
Prevent instances of this class from being copied.
void add_option(std::string option)
Adds option to the existing build option list.
std::string name() const
Returns program name.
void add_options(const StringSet &options)
Appends given build options to the current's objects options.
cl::NDRange default_ndrange() const
Return the default NDRange for the device.
void clear_programs_cache()
Clear the library's cache of binary programs.
void set_device(cl::Device device)
Sets the CL device for which the programs are created.
CLKernelLibrary()
Default Constructor.
std::set< std::string > build_options
std::string kernel_name
std::string get_program_source(const std::string &program_name)
Gets the source of the selected program.
static bool build(const cl::Program &program, const std::string &build_options="")
Build the given CL program.
void add_option_if(bool cond, std::string option)
Adds option if a given condition is true;.
void init(std::string kernel_path, cl::Context context, cl::Device device)
Initialises the kernel library.
Kernel & operator=(const Kernel &)=default
Default copy assignment operator.
Program()
Default constructor.
bool int64_base_atomics_supported() const
Returns true if int64_base_atomics extension is supported by the CL device.
bool fp16_supported() const
Returns true if FP16 is supported by the CL device.
Kernel create_kernel(const std::string &kernel_name, const StringSet &build_options_set={}) const
Creates a kernel from the kernel library.
CLKernelLibrary class.
Program & operator=(const Program &)=default
Default copy assignment operator.
std::string get_kernel_path()
Gets the path that the kernels reside in.
std::string name() const
Returns kernel name.
const std::map< std::string, cl::Program > & get_built_programs() const
Access the cache of built OpenCL programs.
void add_options_if(bool cond, const StringSet &options)
Appends given build options to the current's objects options if a given condition is true.
cl::Context & context()
Accessor for the associated CL context.
void add_option_if_else(bool cond, std::string option_true, std::string option_false)
Adds first option if condition is true else the second one.