openCARP
Doxygen code documentation for the open cardiac electrophysiology simulator openCARP
target.h
Go to the documentation of this file.
1 // ----------------------------------------------------------------------------
2 // openCARP is an open cardiac electrophysiology simulator.
3 //
4 // Copyright (C) 2020 openCARP project
5 //
6 // This program is licensed under the openCARP Academic Public License (APL)
7 // v1.0: You can use and redistribute it and/or modify it in non-commercial
8 // academic environments under the terms of APL as published by the openCARP
9 // project v1.0, or (at your option) any later version. Commercial use requires
10 // a commercial license (info@opencarp.org).
11 //
12 // This program is distributed without any warranty; see the openCARP APL for
13 // more details.
14 //
15 // You should have received a copy of the openCARP APL along with this program
16 // and can find it online: http://www.opencarp.org/license
17 // ----------------------------------------------------------------------------
18 
25 #ifndef TARGETS_H
26 #define TARGETS_H
27 
28 #include <stdexcept>
29 #include <iostream>
30 #include <string>
31 #include <cstring>
32 #ifdef HAS_ROCM_MODEL
33 #include <hip/hip_runtime.h>
34 #endif
35 #ifdef HAS_CUDA_MODEL
36 #include <cuda_runtime.h>
37 #endif
38 
39 namespace limpet
40 {
41 
45 enum Target {
46  AUTO = -2, // !< special value for chosing the target automatically
47  UNKNOWN = -1,
48  CPU,
53 };
54 
62 Target get_target_from_string(std::string const str);
63 
71 std::string get_string_from_target(Target const target);
72 
77 std::string get_target_list_string();
78 
84 bool is_gpu(Target const target);
85 
94 bool is_concrete(Target const target);
95 
103 template <typename T>
105  typedef T value_type;
106 
114  TargetAllocator(Target target, bool always_managed = false) : _target(target), _always_managed(always_managed) {
115  if (!is_concrete(target)) {
116  throw std::invalid_argument("attempting to construct a TargetAllocator with an invalid target");
117  }
118  }
119 
125  Target get_target() const {
126  return this->_target;
127  }
128 
134  void set_target(Target new_target) {
135  if (!is_concrete(new_target)) {
136  throw std::invalid_argument("attempting to set TargetAllocator to an invalid target");
137  }
138  else {
139  this->_target = new_target;
140  }
141  }
142 
154  T* allocate(std::size_t n, bool do_zero = true) {
155  using type = typename std::conditional<std::is_void<T>::value, char, T>::type;
156  T* ptr = nullptr;
157  std::size_t n_bytes = n * sizeof(type);
158  switch (this->_target) {
159  case Target::MLIR_CPU:
160  case Target::CPU: {
161  std::allocator<type> std_alloc;
162  if(n)
163  {
164  ptr = std_alloc.allocate(n);
165  // Explicitely cast the pointer to silence warning on setting a
166  // dynamic class object to 0.
167  // The memory allocated by this function should be initialized after,
168  // for example by using a placement new
169  if(do_zero)
170  std::memset((void *) ptr, 0, n_bytes);
171  }
172 
173 
174  break;
175  }
176  case Target::MLIR_ROCM:
177 #ifdef HAS_ROCM_MODEL
178  {
179  T* hip_ptr;
180  hipError_t error;
181  // Force usage of HIP managed memory even if it's not supported.
182  // This should still preserve the semantics of managed memory, but
183  // is slower. It can be useful to allocate small shared data between CPU
184  // and GPU.
185  if (this->_always_managed) {
186  error = hipMallocManaged(&hip_ptr, n_bytes);
187  }
188  else {
189  int device;
190  int managed_memory = 0;
191  hipGetDevice(&device);
192  hipDeviceGetAttribute(&managed_memory, hipDeviceAttributeManagedMemory, device);
193  // Check for HIP managed memory and fallback to a normal hipMalloc if
194  // it's not supported. It should still allow the CPU to write to GPU memory.
195  if (managed_memory) {
196  error = hipMallocManaged(&hip_ptr, n_bytes);
197  }
198  else {
199  error = hipMalloc(&hip_ptr, n_bytes);
200  }
201  }
202  if (error != hipSuccess) {
203  throw std::runtime_error(hipGetErrorString(error));
204  }
205  error = hipMemset(hip_ptr, 0, n_bytes);
206  if (error != hipSuccess) {
207  throw std::runtime_error(hipGetErrorString(error));
208  }
209  ptr = hip_ptr;
210  }
211 #else
212  throw std::invalid_argument("target MLIR_ROCM is unavailable");
213 #endif
214  break;
215  case Target::MLIR_CUDA:
216 #ifdef HAS_CUDA_MODEL
217  {
218  T* cuda_ptr;
219  cudaError_t error;
220  // Always use unified memory for CUDA
221  error = cudaMallocManaged(&cuda_ptr, n_bytes);
222  if (error != cudaSuccess) {
223  throw std::runtime_error(cudaGetErrorString(error));
224  }
225  error = cudaMemset(cuda_ptr, 0, n_bytes);
226  if (error != cudaSuccess) {
227  throw std::runtime_error(cudaGetErrorString(error));
228  }
229  ptr = cuda_ptr;
230  }
231 #else
232  throw std::invalid_argument("target MLIR_CUDA is unavailable");
233 #endif
234  break;
235  default:
236  throw std::invalid_argument("unknown allocation target");
237  }
238  return ptr;
239  }
240 
241  // Size is not needed
250  void deallocate(T* ptr, std::size_t n = 0) {
251  using type = typename std::conditional<std::is_void<T>::value, char, T>::type;
252  switch (this->_target) {
253  case Target::MLIR_CPU:
254  case Target::CPU: {
255  std::allocator<type> std_alloc;
256  std_alloc.deallocate((type*) ptr, n);
257  }
258  break;
259  case Target::MLIR_ROCM:
260 #ifdef HAS_ROCM_MODEL
261  {
262  hipError_t error;
263  error = hipFree(ptr);
264  if (error != hipSuccess) {
265  throw std::runtime_error(hipGetErrorString(error));
266  }
267  }
268 #else
269  throw std::invalid_argument("target MLIR_ROCM is unavailable");
270 #endif
271  break;
272  case Target::MLIR_CUDA:
273 #ifdef HAS_CUDA_MODEL
274  {
275  cudaError_t error;
276  error = cudaFree(ptr);
277  if (error != cudaSuccess) {
278  throw std::runtime_error(cudaGetErrorString(error));
279  }
280  }
281 #else
282  throw std::invalid_argument("target MLIR_CUDA is unavailable");
283 #endif
284  break;
285  default:
286  throw std::invalid_argument("unknown allocation target");
287  }
288  }
289 
290  private:
291  Target _target;
292  bool _always_managed;
293 };
294 
303 template<typename T>
304 T* allocate_on_target(Target target, std::size_t n, bool always_managed = false, bool do_zero = true) {
305  TargetAllocator<T> alloc(target, always_managed);
306  return alloc.allocate(n, do_zero);
307 }
317 template<typename T>
318 void deallocate_on_target(Target target, T* ptr) {
319  TargetAllocator<T> alloc(target);
320  return alloc.deallocate(ptr);
321 }
322 
323 } // namespace limpet
324 
325 #endif // TARGETS_H
std::string get_string_from_target(Target const target)
Get a string representation of a given target.
Definition: target.cc:46
TargetAllocator(Target target, bool always_managed=false)
Construct a TargetAllocator.
Definition: target.h:114
vectorized CPU code generated with MLIR
Definition: target.h:49
Allocator structure for dynamically allocating memory on multiple targets.
Definition: target.h:104
bool is_gpu(Target const target)
Checks if this is a GPU target.
Definition: target.cc:64
special value to handle unknown targets
Definition: target.h:47
T value_type
type to allocate
Definition: target.h:105
std::string get_target_list_string()
Returns a string containing the list of available targets.
Definition: target.cc:55
a token to indicate the maximum number of targets
Definition: target.h:52
bool is_concrete(Target const target)
Checks if target is a real, concrete target.
Definition: target.cc:68
void set_target(Target new_target)
Set a new target for this allocator.
Definition: target.h:134
Target get_target() const
Get the target for this allocator.
Definition: target.h:125
baseline CPU model generated with the original opencarp code generator
Definition: target.h:48
CUDA code for NVIDIA GPUs generated with MLIR.
Definition: target.h:51
T * allocate(std::size_t n, bool do_zero=true)
Allocate memory for type T.
Definition: target.h:154
Target get_target_from_string(std::string const str)
Returns a value from the Target enum from a given string.
Definition: target.cc:36
void deallocate(T *ptr, std::size_t n=0)
Deallocate memory pointed by ptr.
Definition: target.h:250
Target
enum that represents different targets to run ionic models on.
Definition: target.h:45
T * allocate_on_target(Target target, std::size_t n, bool always_managed=false, bool do_zero=true)
Utility function for allocating memory on a target. See TargetAllocator.
Definition: target.h:304
void deallocate_on_target(Target target, T *ptr)
Utility function for deallocating memory on a target. See TargetAllocator.
Definition: target.h:318
ROCM code for AMD GPUs generated with MLIR.
Definition: target.h:50