OmniSciDB  d2f719934e
 All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Macros Groups Pages
L0Mgr.cpp
Go to the documentation of this file.
1 /*
2  * Copyright 2020 OmniSci, Inc.
3  *
4  * Licensed under the Apache License, Version 2.0 (the "License");
5  * you may not use this file except in compliance with the License.
6  * You may obtain a copy of the License at
7  *
8  * http://www.apache.org/licenses/LICENSE-2.0
9  *
10  * Unless required by applicable law or agreed to in writing, software
11  * distributed under the License is distributed on an "AS IS" BASIS,
12  * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13  * See the License for the specific language governing permissions and
14  * limitations under the License.
15  */
16 
17 #include "L0Mgr/L0Mgr.h"
18 
19 #include "Logger/Logger.h"
20 #include "Utils.h"
21 
22 #include <iostream>
23 #include <limits>
24 
25 #include <level_zero/ze_api.h>
26 
27 namespace l0 {
28 
29 L0Driver::L0Driver(ze_driver_handle_t handle) : driver_(handle) {
30  ze_context_desc_t ctx_desc = {ZE_STRUCTURE_TYPE_CONTEXT_DESC, nullptr, 0};
31  L0_SAFE_CALL(zeContextCreate(driver_, &ctx_desc, &context_));
32 
33  uint32_t device_count = 0;
34  L0_SAFE_CALL(zeDeviceGet(driver_, &device_count, nullptr));
35 
36  std::vector<ze_device_handle_t> devices(device_count);
37  L0_SAFE_CALL(zeDeviceGet(driver_, &device_count, devices.data()));
38 
39  for (auto device : devices) {
40  ze_device_properties_t device_properties;
41  L0_SAFE_CALL(zeDeviceGetProperties(device, &device_properties));
42  if (ZE_DEVICE_TYPE_GPU == device_properties.type) {
43  devices_.push_back(std::make_shared<L0Device>(*this, device));
44  }
45  }
46 }
47 
48 L0Driver::~L0Driver() {
49  auto status = (zeContextDestroy(context_));
50  if (status) {
51  std::cerr << "Non-zero status for context destructor" << std::endl;
52  }
53 }
54 
55 ze_context_handle_t L0Driver::ctx() const {
56  return context_;
57 }
58 
59 ze_driver_handle_t L0Driver::driver() const {
60  return driver_;
61 }
62 
63 const std::vector<std::shared_ptr<L0Device>>& L0Driver::devices() const {
64  return devices_;
65 }
66 
67 std::vector<std::shared_ptr<L0Driver>> get_drivers() {
68  zeInit(0);
69  uint32_t driver_count = 0;
70  zeDriverGet(&driver_count, nullptr);
71 
72  std::vector<ze_driver_handle_t> handles(driver_count);
73  zeDriverGet(&driver_count, handles.data());
74 
75  std::vector<std::shared_ptr<L0Driver>> result(driver_count);
76  for (int i = 0; i < driver_count; i++) {
77  result[i] = std::make_shared<L0Driver>(handles[i]);
78  }
79  return result;
80 }
81 
82 L0CommandList::L0CommandList(ze_command_list_handle_t handle) : handle_(handle) {}
83 
84 void L0CommandList::submit(L0CommandQueue& queue) {
85  L0_SAFE_CALL(zeCommandListClose(handle_));
86  L0_SAFE_CALL(zeCommandQueueExecuteCommandLists(queue.handle(), 1, &handle_, nullptr));
88  zeCommandQueueSynchronize(queue.handle(), std::numeric_limits<uint32_t>::max()));
89 }
90 
91 ze_command_list_handle_t L0CommandList::handle() const {
92  return handle_;
93 }
94 
95 L0CommandList::~L0CommandList() {
96  // TODO: maybe return to pool
97 }
98 
99 void L0CommandList::copy(void* dst, const void* src, const size_t num_bytes) {
100  L0_SAFE_CALL(
101  zeCommandListAppendMemoryCopy(handle_, dst, src, num_bytes, nullptr, 0, nullptr));
102  L0_SAFE_CALL(zeCommandListAppendBarrier(handle_, nullptr, 0, nullptr));
103 }
104 
105 void* allocate_device_mem(const size_t num_bytes, L0Device& device) {
106  ze_device_mem_alloc_desc_t alloc_desc;
107  alloc_desc.stype = ZE_STRUCTURE_TYPE_DEVICE_MEM_ALLOC_DESC;
108  alloc_desc.pNext = nullptr;
109  alloc_desc.flags = 0;
110  alloc_desc.ordinal = 0;
111 
112  void* mem;
113  L0_SAFE_CALL(zeMemAllocDevice(
114  device.ctx(), &alloc_desc, num_bytes, 0 /*align*/, device.device(), &mem));
115  return mem;
116 }
117 
118 L0Device::L0Device(const L0Driver& driver, ze_device_handle_t device)
119  : device_(device), driver_(driver) {
120  ze_command_queue_handle_t queue_handle;
121  ze_command_queue_desc_t command_queue_desc = {ZE_STRUCTURE_TYPE_COMMAND_QUEUE_DESC,
122  nullptr,
123  0,
124  0,
125  0,
126  ZE_COMMAND_QUEUE_MODE_DEFAULT,
127  ZE_COMMAND_QUEUE_PRIORITY_NORMAL};
128  L0_SAFE_CALL(
129  zeCommandQueueCreate(driver_.ctx(), device_, &command_queue_desc, &queue_handle));
130 
131  command_queue_ = std::make_shared<L0CommandQueue>(queue_handle);
132 }
133 
134 L0Device::~L0Device() {}
135 
136 ze_context_handle_t L0Device::ctx() const {
137  return driver_.ctx();
138 }
139 ze_device_handle_t L0Device::device() const {
140  return device_;
141 }
142 std::shared_ptr<L0CommandQueue> L0Device::command_queue() const {
143  return command_queue_;
144 }
145 
146 std::unique_ptr<L0CommandList> L0Device::create_command_list() const {
147  ze_command_list_desc_t desc = {
148  ZE_STRUCTURE_TYPE_COMMAND_LIST_DESC,
149  nullptr,
150  0,
151  0 // flags
152  };
153  ze_command_list_handle_t res;
154  zeCommandListCreate(ctx(), device_, &desc, &res);
155  return std::make_unique<L0CommandList>(res);
156 }
157 
158 L0CommandQueue::L0CommandQueue(ze_command_queue_handle_t handle) : handle_(handle) {}
159 
160 ze_command_queue_handle_t L0CommandQueue::handle() const {
161  return handle_;
162 }
163 
164 L0CommandQueue::~L0CommandQueue() {
165  auto status = (zeCommandQueueDestroy(handle_));
166  if (status) {
167  std::cerr << "Non-zero status for command queue destructor" << std::endl;
168  }
169 }
170 
171 std::shared_ptr<L0Module> L0Device::create_module(uint8_t* code,
172  size_t len,
173  bool log) const {
174  ze_module_desc_t desc{
175  .stype = ZE_STRUCTURE_TYPE_MODULE_DESC,
176  .pNext = nullptr,
177  .format = ZE_MODULE_FORMAT_IL_SPIRV,
178  .inputSize = len,
179  .pInputModule = code,
180  .pBuildFlags = "",
181  .pConstants = nullptr,
182  };
183  ze_module_handle_t handle;
184  ze_module_build_log_handle_t buildlog = nullptr;
185 
186  auto status = zeModuleCreate(ctx(), device_, &desc, &handle, &buildlog);
187  if (log) {
188  size_t logSize = 0;
189  L0_SAFE_CALL(zeModuleBuildLogGetString(buildlog, &logSize, nullptr));
190  std::vector<char> strLog(logSize);
191  L0_SAFE_CALL(zeModuleBuildLogGetString(buildlog, &logSize, strLog.data()));
192  std::fstream out;
193  out.open("log.txt", std::ios::app);
194  if (!out.good()) {
195  std::cerr << "Unable to open log file.\n";
196  } else {
197  out << std::string(strLog.begin(), strLog.end());
198  out.close();
199  }
200  }
201  if (status) {
202  throw l0::L0Exception(status);
203  }
204  return std::make_shared<L0Module>(handle);
205 }
206 
208 
209 const std::vector<std::shared_ptr<L0Driver>>& L0Manager::drivers() const {
210  return drivers_;
211 }
212 
213 int8_t* L0Manager::allocateDeviceMem(const size_t num_bytes, int device_id) {
214  auto& device = drivers_[0]->devices()[device_id];
215  return (int8_t*)allocate_device_mem(num_bytes, *device);
216 }
217 
218 L0Module::L0Module(ze_module_handle_t handle) : handle_(handle) {}
219 
220 ze_module_handle_t L0Module::handle() const {
221  return handle_;
222 }
223 
224 L0Module::~L0Module() {
225  auto status = zeModuleDestroy(handle_);
226  if (status) {
227  std::cerr << "Non-zero status for command module destructor" << std::endl;
228  }
229 }
230 
231 std::shared_ptr<L0Kernel> L0Module::create_kernel(const char* name,
232  uint32_t x,
233  uint32_t y,
234  uint32_t z) const {
235  ze_kernel_desc_t desc{
236  .stype = ZE_STRUCTURE_TYPE_KERNEL_DESC,
237  .pNext = nullptr,
238  .flags = 0,
239  .pKernelName = name,
240  };
241  ze_kernel_handle_t handle;
242  L0_SAFE_CALL(zeKernelCreate(this->handle_, &desc, &handle));
243  return std::make_shared<L0Kernel>(handle, x, y, z);
244 }
245 
246 L0Kernel::L0Kernel(ze_kernel_handle_t handle, uint32_t x, uint32_t y, uint32_t z)
247  : handle_(handle), group_size_({x, y, z}) {
248  zeKernelSetGroupSize(handle_, x, y, z);
249 }
250 
251 ze_group_count_t& L0Kernel::group_size() {
252  return group_size_;
253 }
254 
255 ze_kernel_handle_t L0Kernel::handle() const {
256  return handle_;
257 }
258 
259 L0Kernel::~L0Kernel() {
260  auto status = zeKernelDestroy(handle_);
261  if (status) {
262  std::cerr << "Non-zero status for command kernel destructor" << std::endl;
263  }
264 }
265 
266 void L0Manager::copyHostToDevice(int8_t* device_ptr,
267  const int8_t* host_ptr,
268  const size_t num_bytes,
269  const int device_num) {
270  auto& device = drivers()[0]->devices()[device_num];
271  auto cl = device->create_command_list();
272  auto queue = device->command_queue();
273 
274  cl->copy(device_ptr, host_ptr, num_bytes);
275  cl->submit(*queue);
276 }
277 
278 void L0Manager::copyDeviceToHost(int8_t* host_ptr,
279  const int8_t* device_ptr,
280  const size_t num_bytes,
281  const int device_num) {
282  auto& device = drivers_[0]->devices()[device_num];
283  auto cl = device->create_command_list();
284  auto queue = device->command_queue();
285 
286  cl->copy(host_ptr, device_ptr, num_bytes);
287  cl->submit(*queue);
288 }
289 
290 void L0Manager::copyDeviceToDevice(int8_t* dest_ptr,
291  int8_t* src_ptr,
292  const size_t num_bytes,
293  const int dest_device_num,
294  const int src_device_num) {
295  CHECK(false);
296 }
297 
298 int8_t* L0Manager::allocatePinnedHostMem(const size_t num_bytes) {
299  CHECK(false);
300  return nullptr;
301 }
302 
303 void L0Manager::freePinnedHostMem(int8_t* host_ptr) {
304  CHECK(false);
305 }
306 
307 void L0Manager::freeDeviceMem(int8_t* device_ptr) {
308  auto ctx = drivers_[0]->ctx();
309  L0_SAFE_CALL(zeMemFree(ctx, device_ptr));
310 }
311 
312 void L0Manager::zeroDeviceMem(int8_t* device_ptr,
313  const size_t num_bytes,
314  const int device_num) {
315  setDeviceMem(device_ptr, 0, num_bytes, device_num);
316 }
317 void L0Manager::setDeviceMem(int8_t* device_ptr,
318  const unsigned char uc,
319  const size_t num_bytes,
320  const int device_num) {
321  auto& device = drivers_[0]->devices()[device_num];
322  auto cl = device->create_command_list();
323  L0_SAFE_CALL(zeCommandListAppendMemoryFill(
324  cl->handle(), device_ptr, &uc, 1, num_bytes, nullptr, 0, nullptr));
325  cl->submit(*device->command_queue());
326 }
327 
329  for (auto& device : drivers_[0]->devices()) {
330  L0_SAFE_CALL(zeCommandQueueSynchronize(device->command_queue()->handle(),
331  std::numeric_limits<uint32_t>::max()));
332  }
333 }
334 } // namespace l0
void freeDeviceMem(int8_t *device_ptr)
Definition: L0Mgr.cpp:307
void zeroDeviceMem(int8_t *device_ptr, const size_t num_bytes, const int device_num)
Definition: L0Mgr.cpp:312
#define L0_SAFE_CALL(call)
Definition: Utils.h:19
void copyHostToDevice(int8_t *device_ptr, const int8_t *host_ptr, const size_t num_bytes, const int device_num)
Definition: L0Mgr.cpp:266
string name
Definition: setup.in.py:72
void setDeviceMem(int8_t *device_ptr, const unsigned char uc, const size_t num_bytes, const int device_num)
Definition: L0Mgr.cpp:317
std::shared_ptr< L0Module > create_module(uint8_t *code, size_t len, bool log=false) const
Definition: L0Mgr.cpp:171
std::vector< std::shared_ptr< L0Driver > > get_drivers()
Definition: L0Mgr.cpp:67
DEVICE auto copy(ARGS &&...args)
Definition: gpu_enabled.h:51
void copyDeviceToDevice(int8_t *dest_ptr, int8_t *src_ptr, const size_t num_bytes, const int dest_device_num, const int src_device_num)
Definition: L0Mgr.cpp:290
std::vector< std::shared_ptr< L0Device > > devices_
Definition: L0Mgr.h:34
void synchronizeDevices() const
Definition: L0Mgr.cpp:328
void copyDeviceToHost(int8_t *host_ptr, const int8_t *device_ptr, const size_t num_bytes, const int device_num)
Definition: L0Mgr.cpp:278
void freePinnedHostMem(int8_t *host_ptr)
Definition: L0Mgr.cpp:303
const std::vector< std::shared_ptr< L0Driver > > & drivers() const
Definition: L0Mgr.cpp:209
#define CHECK(condition)
Definition: Logger.h:211
int8_t * allocatePinnedHostMem(const size_t num_bytes)
Definition: L0Mgr.cpp:298
std::vector< std::shared_ptr< L0Driver > > drivers_
Definition: L0Mgr.h:209
void * allocate_device_mem(const size_t num_bytes, L0Device &device)
Definition: L0Mgr.cpp:105
int8_t * allocateDeviceMem(const size_t num_bytes, const int device_num)
Definition: L0Mgr.cpp:213
const std::vector< std::shared_ptr< L0Device > > & devices() const
Definition: L0Mgr.cpp:63
std::shared_ptr< L0Kernel > create_kernel(const char *name, uint32_t x, uint32_t y, uint32_t z) const
Definition: L0Mgr.cpp:231