ViennaCL - The Vienna Computing Library  1.7.0
Free open-source GPU-accelerated linear algebra and solver library.
memory.hpp
Go to the documentation of this file.
1 #ifndef VIENNACL_BACKEND_MEMORY_HPP
2 #define VIENNACL_BACKEND_MEMORY_HPP
3 
4 /* =========================================================================
5  Copyright (c) 2010-2015, Institute for Microelectronics,
6  Institute for Analysis and Scientific Computing,
7  TU Wien.
8  Portions of this software are copyright by UChicago Argonne, LLC.
9 
10  -----------------
11  ViennaCL - The Vienna Computing Library
12  -----------------
13 
14  Project Head: Karl Rupp rupp@iue.tuwien.ac.at
15 
16  (A list of authors and contributors can be found in the manual)
17 
18  License: MIT (X11), see file LICENSE in the base directory
19 ============================================================================= */
20 
25 #include <vector>
26 #include <cassert>
27 #include "viennacl/forwards.h"
29 #include "viennacl/context.hpp"
33 
35 
36 #ifdef VIENNACL_WITH_OPENCL
38 #include "viennacl/ocl/backend.hpp"
39 #endif
40 
41 #ifdef VIENNACL_WITH_CUDA
43 #endif
44 
45 
46 namespace viennacl
47 {
48 namespace backend
49 {
50 
51 
52  // if a user compiles with CUDA, it is reasonable to expect that CUDA should be the default
54  inline void finish()
55  {
56 #ifdef VIENNACL_WITH_CUDA
57  cudaDeviceSynchronize();
58 #endif
59 #ifdef VIENNACL_WITH_OPENCL
61 #endif
62  }
63 
64 
65 
66 
67  // Requirements for backend:
68 
69  // ---- Memory ----
70  //
71  // * memory_create(size, host_ptr)
72  // * memory_copy(src, dest, offset_src, offset_dest, size)
73  // * memory_write(src, offset, size, ptr)
74  // * memory_read(src, offset, size, ptr)
75  //
76 
87  inline void memory_create(mem_handle & handle, vcl_size_t size_in_bytes, viennacl::context const & ctx, const void * host_ptr = NULL)
88  {
89  if (size_in_bytes > 0)
90  {
93 
94  switch (handle.get_active_handle_id())
95  {
96  case MAIN_MEMORY:
97  handle.ram_handle() = cpu_ram::memory_create(size_in_bytes, host_ptr);
98  handle.raw_size(size_in_bytes);
99  break;
100 #ifdef VIENNACL_WITH_OPENCL
101  case OPENCL_MEMORY:
102  handle.opencl_handle().context(ctx.opencl_context());
103  handle.opencl_handle() = opencl::memory_create(handle.opencl_handle().context(), size_in_bytes, host_ptr);
104  handle.raw_size(size_in_bytes);
105  break;
106 #endif
107 #ifdef VIENNACL_WITH_CUDA
108  case CUDA_MEMORY:
109  handle.cuda_handle() = cuda::memory_create(size_in_bytes, host_ptr);
110  handle.raw_size(size_in_bytes);
111  break;
112 #endif
114  throw memory_exception("not initialised!");
115  default:
116  throw memory_exception("unknown memory handle!");
117  }
118  }
119  }
120 
121  /*
122  inline void memory_create(mem_handle & handle, vcl_size_t size_in_bytes, const void * host_ptr = NULL)
123  {
124  viennacl::context ctx(default_memory_type());
125  memory_create(handle, size_in_bytes, ctx, host_ptr);
126  }*/
127 
128 
140  inline void memory_copy(mem_handle const & src_buffer,
141  mem_handle & dst_buffer,
142  vcl_size_t src_offset,
143  vcl_size_t dst_offset,
144  vcl_size_t bytes_to_copy)
145  {
146  assert( src_buffer.get_active_handle_id() == dst_buffer.get_active_handle_id() && bool("memory_copy() must be called on buffers from the same domain") );
147 
148  if (bytes_to_copy > 0)
149  {
150  switch (src_buffer.get_active_handle_id())
151  {
152  case MAIN_MEMORY:
153  cpu_ram::memory_copy(src_buffer.ram_handle(), dst_buffer.ram_handle(), src_offset, dst_offset, bytes_to_copy);
154  break;
155 #ifdef VIENNACL_WITH_OPENCL
156  case OPENCL_MEMORY:
157  opencl::memory_copy(src_buffer.opencl_handle(), dst_buffer.opencl_handle(), src_offset, dst_offset, bytes_to_copy);
158  break;
159 #endif
160 #ifdef VIENNACL_WITH_CUDA
161  case CUDA_MEMORY:
162  cuda::memory_copy(src_buffer.cuda_handle(), dst_buffer.cuda_handle(), src_offset, dst_offset, bytes_to_copy);
163  break;
164 #endif
166  throw memory_exception("not initialised!");
167  default:
168  throw memory_exception("unknown memory handle!");
169  }
170  }
171  }
172 
173  // TODO: Refine this concept. Maybe move to constructor?
177  inline void memory_shallow_copy(mem_handle const & src_buffer,
178  mem_handle & dst_buffer)
179  {
180  assert( (dst_buffer.get_active_handle_id() == MEMORY_NOT_INITIALIZED) && bool("Shallow copy on already initialized memory not supported!"));
181 
182  switch (src_buffer.get_active_handle_id())
183  {
184  case MAIN_MEMORY:
185  dst_buffer.switch_active_handle_id(src_buffer.get_active_handle_id());
186  dst_buffer.ram_handle() = src_buffer.ram_handle();
187  dst_buffer.raw_size(src_buffer.raw_size());
188  break;
189 #ifdef VIENNACL_WITH_OPENCL
190  case OPENCL_MEMORY:
191  dst_buffer.switch_active_handle_id(src_buffer.get_active_handle_id());
192  dst_buffer.opencl_handle() = src_buffer.opencl_handle();
193  dst_buffer.raw_size(src_buffer.raw_size());
194  break;
195 #endif
196 #ifdef VIENNACL_WITH_CUDA
197  case CUDA_MEMORY:
198  dst_buffer.switch_active_handle_id(src_buffer.get_active_handle_id());
199  dst_buffer.cuda_handle() = src_buffer.cuda_handle();
200  dst_buffer.raw_size(src_buffer.raw_size());
201  break;
202 #endif
204  throw memory_exception("not initialised!");
205  default:
206  throw memory_exception("unknown memory handle!");
207  }
208  }
209 
220  inline void memory_write(mem_handle & dst_buffer,
221  vcl_size_t dst_offset,
222  vcl_size_t bytes_to_write,
223  const void * ptr,
224  bool async = false)
225  {
226  if (bytes_to_write > 0)
227  {
228  switch (dst_buffer.get_active_handle_id())
229  {
230  case MAIN_MEMORY:
231  cpu_ram::memory_write(dst_buffer.ram_handle(), dst_offset, bytes_to_write, ptr, async);
232  break;
233 #ifdef VIENNACL_WITH_OPENCL
234  case OPENCL_MEMORY:
235  opencl::memory_write(dst_buffer.opencl_handle(), dst_offset, bytes_to_write, ptr, async);
236  break;
237 #endif
238 #ifdef VIENNACL_WITH_CUDA
239  case CUDA_MEMORY:
240  cuda::memory_write(dst_buffer.cuda_handle(), dst_offset, bytes_to_write, ptr, async);
241  break;
242 #endif
244  throw memory_exception("not initialised!");
245  default:
246  throw memory_exception("unknown memory handle!");
247  }
248  }
249  }
250 
261  inline void memory_read(mem_handle const & src_buffer,
262  vcl_size_t src_offset,
263  vcl_size_t bytes_to_read,
264  void * ptr,
265  bool async = false)
266  {
267  //finish(); //Fixes some issues with AMD APP SDK. However, might sacrifice a few percents of performance in some cases.
268 
269  if (bytes_to_read > 0)
270  {
271  switch (src_buffer.get_active_handle_id())
272  {
273  case MAIN_MEMORY:
274  cpu_ram::memory_read(src_buffer.ram_handle(), src_offset, bytes_to_read, ptr, async);
275  break;
276 #ifdef VIENNACL_WITH_OPENCL
277  case OPENCL_MEMORY:
278  opencl::memory_read(src_buffer.opencl_handle(), src_offset, bytes_to_read, ptr, async);
279  break;
280 #endif
281 #ifdef VIENNACL_WITH_CUDA
282  case CUDA_MEMORY:
283  cuda::memory_read(src_buffer.cuda_handle(), src_offset, bytes_to_read, ptr, async);
284  break;
285 #endif
287  throw memory_exception("not initialised!");
288  default:
289  throw memory_exception("unknown memory handle!");
290  }
291  }
292  }
293 
294 
295 
296  namespace detail
297  {
298  template<typename T>
300  {
301  return sizeof(T);
302  }
303 
304 
305  template<>
307  #ifdef VIENNACL_WITH_OPENCL
308  mem_type //in order to compile cleanly at -Wextra in GCC
309  #endif
310  )
311  {
312 #ifdef VIENNACL_WITH_OPENCL
313  if (mem_type == OPENCL_MEMORY)
314  return sizeof(cl_ulong);
315 #endif
316  return sizeof(unsigned long);
317  }
318 
319  template<>
321  #ifdef VIENNACL_WITH_OPENCL
322  mem_type //in order to compile cleanly at -Wextra in GCC
323  #endif
324  )
325  {
326 #ifdef VIENNACL_WITH_OPENCL
327  if (mem_type == OPENCL_MEMORY)
328  return sizeof(cl_long);
329 #endif
330  return sizeof(long);
331  }
332 
333 
334  template<>
336  #ifdef VIENNACL_WITH_OPENCL
337  mem_type //in order to compile cleanly at -Wextra in GCC
338  #endif
339  )
340  {
341 #ifdef VIENNACL_WITH_OPENCL
342  if (mem_type == OPENCL_MEMORY)
343  return sizeof(cl_uint);
344 #endif
345  return sizeof(unsigned int);
346  }
347 
348  template<>
350  #ifdef VIENNACL_WITH_OPENCL
351  mem_type //in order to compile cleanly at -Wextra in GCC
352  #endif
353  )
354  {
355 #ifdef VIENNACL_WITH_OPENCL
356  if (mem_type == OPENCL_MEMORY)
357  return sizeof(cl_int);
358 #endif
359  return sizeof(int);
360  }
361 
362 
363  }
364 
365 
367  template<typename DataType>
369  {
370  if (handle.get_active_handle_id() == new_ctx.memory_type())
371  return;
372 
373  if (handle.get_active_handle_id() == viennacl::MEMORY_NOT_INITIALIZED || handle.raw_size() == 0)
374  {
375  handle.switch_active_handle_id(new_ctx.memory_type());
376 #ifdef VIENNACL_WITH_OPENCL
377  if (new_ctx.memory_type() == OPENCL_MEMORY)
378  handle.opencl_handle().context(new_ctx.opencl_context());
379 #endif
380  return;
381  }
382 
383  vcl_size_t size_dst = detail::element_size<DataType>(handle.get_active_handle_id());
384  vcl_size_t size_src = detail::element_size<DataType>(new_ctx.memory_type());
385 
386  if (size_dst != size_src) // OpenCL data element size not the same as host data element size
387  {
388  throw memory_exception("Heterogeneous data element sizes not yet supported!");
389  }
390  else //no data conversion required
391  {
392  if (handle.get_active_handle_id() == MAIN_MEMORY) //we can access the existing data directly
393  {
394  switch (new_ctx.memory_type())
395  {
396 #ifdef VIENNACL_WITH_OPENCL
397  case OPENCL_MEMORY:
398  handle.opencl_handle().context(new_ctx.opencl_context());
399  handle.opencl_handle() = opencl::memory_create(handle.opencl_handle().context(), handle.raw_size(), handle.ram_handle().get());
400  break;
401 #endif
402 #ifdef VIENNACL_WITH_CUDA
403  case CUDA_MEMORY:
404  handle.cuda_handle() = cuda::memory_create(handle.raw_size(), handle.ram_handle().get());
405  break;
406 #endif
407  case MAIN_MEMORY:
408  default:
409  throw memory_exception("Invalid destination domain");
410  }
411  }
412 #ifdef VIENNACL_WITH_OPENCL
413  else if (handle.get_active_handle_id() == OPENCL_MEMORY) // data can be dumped into destination directly
414  {
415  std::vector<DataType> buffer;
416 
417  switch (new_ctx.memory_type())
418  {
419  case MAIN_MEMORY:
420  handle.ram_handle() = cpu_ram::memory_create(handle.raw_size());
421  opencl::memory_read(handle.opencl_handle(), 0, handle.raw_size(), handle.ram_handle().get());
422  break;
423 #ifdef VIENNACL_WITH_CUDA
424  case CUDA_MEMORY:
425  buffer.resize(handle.raw_size() / sizeof(DataType));
426  opencl::memory_read(handle.opencl_handle(), 0, handle.raw_size(), &(buffer[0]));
427  cuda::memory_create(handle.cuda_handle(), handle.raw_size(), &(buffer[0]));
428  break;
429 #endif
430  default:
431  throw memory_exception("Invalid destination domain");
432  }
433  }
434 #endif
435 #ifdef VIENNACL_WITH_CUDA
436  else //CUDA_MEMORY
437  {
438  std::vector<DataType> buffer;
439 
440  // write
441  switch (new_ctx.memory_type())
442  {
443  case MAIN_MEMORY:
444  handle.ram_handle() = cpu_ram::memory_create(handle.raw_size());
445  cuda::memory_read(handle.cuda_handle(), 0, handle.raw_size(), handle.ram_handle().get());
446  break;
447 #ifdef VIENNACL_WITH_OPENCL
448  case OPENCL_MEMORY:
449  buffer.resize(handle.raw_size() / sizeof(DataType));
450  cuda::memory_read(handle.cuda_handle(), 0, handle.raw_size(), &(buffer[0]));
451  handle.opencl_handle() = opencl::memory_create(handle.raw_size(), &(buffer[0]));
452  break;
453 #endif
454  default:
455  throw memory_exception("Unsupported source memory domain");
456  }
457  }
458 #endif
459 
460  // everything succeeded so far, now switch to new domain:
461  handle.switch_active_handle_id(new_ctx.memory_type());
462 
463  } // no data conversion
464  }
465 
466 
467 
469  template<typename DataType>
470  void typesafe_memory_copy(mem_handle const & handle_src, mem_handle & handle_dst)
471  {
472  if (handle_dst.get_active_handle_id() == MEMORY_NOT_INITIALIZED)
474 
475  vcl_size_t element_size_src = detail::element_size<DataType>(handle_src.get_active_handle_id());
476  vcl_size_t element_size_dst = detail::element_size<DataType>(handle_dst.get_active_handle_id());
477 
478  if (element_size_src != element_size_dst)
479  {
480  // Data needs to be converted.
481 
482  typesafe_host_array<DataType> buffer_src(handle_src);
483  typesafe_host_array<DataType> buffer_dst(handle_dst, handle_src.raw_size() / element_size_src);
484 
485  //
486  // Step 1: Fill buffer_dst depending on where the data resides:
487  //
488  DataType const * src_data;
489  switch (handle_src.get_active_handle_id())
490  {
491  case MAIN_MEMORY:
492  src_data = reinterpret_cast<DataType const *>(handle_src.ram_handle().get());
493  for (vcl_size_t i=0; i<buffer_dst.size(); ++i)
494  buffer_dst.set(i, src_data[i]);
495  break;
496 
497 #ifdef VIENNACL_WITH_OPENCL
498  case OPENCL_MEMORY:
499  buffer_src.resize(handle_src, handle_src.raw_size() / element_size_src);
500  opencl::memory_read(handle_src.opencl_handle(), 0, buffer_src.raw_size(), buffer_src.get());
501  for (vcl_size_t i=0; i<buffer_dst.size(); ++i)
502  buffer_dst.set(i, buffer_src[i]);
503  break;
504 #endif
505 #ifdef VIENNACL_WITH_CUDA
506  case CUDA_MEMORY:
507  buffer_src.resize(handle_src, handle_src.raw_size() / element_size_src);
508  cuda::memory_read(handle_src.cuda_handle(), 0, buffer_src.raw_size(), buffer_src.get());
509  for (vcl_size_t i=0; i<buffer_dst.size(); ++i)
510  buffer_dst.set(i, buffer_src[i]);
511  break;
512 #endif
513 
514  default:
515  throw memory_exception("unsupported memory domain");
516  }
517 
518  //
519  // Step 2: Write to destination
520  //
521  if (handle_dst.raw_size() == buffer_dst.raw_size())
522  viennacl::backend::memory_write(handle_dst, 0, buffer_dst.raw_size(), buffer_dst.get());
523  else
524  viennacl::backend::memory_create(handle_dst, buffer_dst.raw_size(), viennacl::traits::context(handle_dst), buffer_dst.get());
525 
526  }
527  else
528  {
529  // No data conversion required.
530  typesafe_host_array<DataType> buffer(handle_src);
531 
532  switch (handle_src.get_active_handle_id())
533  {
534  case MAIN_MEMORY:
535  switch (handle_dst.get_active_handle_id())
536  {
537  case MAIN_MEMORY:
538  case OPENCL_MEMORY:
539  case CUDA_MEMORY:
540  if (handle_dst.raw_size() == handle_src.raw_size())
541  viennacl::backend::memory_write(handle_dst, 0, handle_src.raw_size(), handle_src.ram_handle().get());
542  else
543  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst), handle_src.ram_handle().get());
544  break;
545 
546  default:
547  throw memory_exception("unsupported destination memory domain");
548  }
549  break;
550 
551  case OPENCL_MEMORY:
552  switch (handle_dst.get_active_handle_id())
553  {
554  case MAIN_MEMORY:
555  if (handle_dst.raw_size() != handle_src.raw_size())
556  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
557  viennacl::backend::memory_read(handle_src, 0, handle_src.raw_size(), handle_dst.ram_handle().get());
558  break;
559 
560  case OPENCL_MEMORY:
561  if (handle_dst.raw_size() != handle_src.raw_size())
562  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
563  viennacl::backend::memory_copy(handle_src, handle_dst, 0, 0, handle_src.raw_size());
564  break;
565 
566  case CUDA_MEMORY:
567  if (handle_dst.raw_size() != handle_src.raw_size())
568  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
569  buffer.resize(handle_src, handle_src.raw_size() / element_size_src);
570  viennacl::backend::memory_read(handle_src, 0, handle_src.raw_size(), buffer.get());
571  viennacl::backend::memory_write(handle_dst, 0, handle_src.raw_size(), buffer.get());
572  break;
573 
574  default:
575  throw memory_exception("unsupported destination memory domain");
576  }
577  break;
578 
579  case CUDA_MEMORY:
580  switch (handle_dst.get_active_handle_id())
581  {
582  case MAIN_MEMORY:
583  if (handle_dst.raw_size() != handle_src.raw_size())
584  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
585  viennacl::backend::memory_read(handle_src, 0, handle_src.raw_size(), handle_dst.ram_handle().get());
586  break;
587 
588  case OPENCL_MEMORY:
589  if (handle_dst.raw_size() != handle_src.raw_size())
590  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
591  buffer.resize(handle_src, handle_src.raw_size() / element_size_src);
592  viennacl::backend::memory_read(handle_src, 0, handle_src.raw_size(), buffer.get());
593  viennacl::backend::memory_write(handle_dst, 0, handle_src.raw_size(), buffer.get());
594  break;
595 
596  case CUDA_MEMORY:
597  if (handle_dst.raw_size() != handle_src.raw_size())
598  viennacl::backend::memory_create(handle_dst, handle_src.raw_size(), viennacl::traits::context(handle_dst));
599  viennacl::backend::memory_copy(handle_src, handle_dst, 0, 0, handle_src.raw_size());
600  break;
601 
602  default:
603  throw memory_exception("unsupported destination memory domain");
604  }
605  break;
606 
607  default:
608  throw memory_exception("unsupported source memory domain");
609  }
610 
611  }
612  }
613 
614 
615 } //backend
616 
617 //
618 // Convenience layer:
619 //
621 template<typename T>
623 {
624  obj.switch_memory_context(new_ctx);
625 }
626 
627 } //viennacl
628 #endif
void typesafe_memory_copy(mem_handle const &handle_src, mem_handle &handle_dst)
Copies data of the provided 'DataType' from 'handle_src' to 'handle_dst' and converts the data if the...
Definition: memory.hpp:470
Helper class implementing an array on the host. Default case: No conversion necessary.
Definition: util.hpp:92
void memory_read(handle_type const &src_buffer, vcl_size_t src_offset, vcl_size_t bytes_to_copy, void *ptr, bool)
Reads data from a buffer back to main RAM.
Definition: cpu_ram.hpp:153
cl_mem memory_create(viennacl::ocl::context const &ctx, vcl_size_t size_in_bytes, const void *host_ptr=NULL)
Creates an array of the specified size in the current OpenCL context. If the second argument is provi...
Definition: opencl.hpp:55
void memory_write(mem_handle &dst_buffer, vcl_size_t dst_offset, vcl_size_t bytes_to_write, const void *ptr, bool async=false)
Writes data from main RAM identified by 'ptr' to the buffer identified by 'dst_buffer'.
Definition: memory.hpp:220
handle_type memory_create(vcl_size_t size_in_bytes, const void *host_ptr=NULL)
Creates an array of the specified size in main RAM. If the second argument is provided, the buffer is initialized with data from that pointer.
Definition: cpu_ram.hpp:73
void finish() const
Waits until all kernels in the queue have finished their execution.
Exception class in case of memory errors.
Definition: forwards.h:572
void memory_write(viennacl::ocl::handle< cl_mem > &dst_buffer, vcl_size_t dst_offset, vcl_size_t bytes_to_copy, const void *ptr, bool async=false)
Writes data from main RAM identified by 'ptr' to the OpenCL buffer identified by 'dst_buffer'.
Definition: opencl.hpp:97
void memory_write(handle_type &dst_buffer, vcl_size_t dst_offset, vcl_size_t bytes_to_copy, const void *ptr, bool async=false)
Writes data from main RAM identified by 'ptr' to the CUDA buffer identified by 'dst_buffer'.
Definition: cuda.hpp:154
void finish()
Synchronizes the execution. finish() will only return after all compute kernels (CUDA, OpenCL) have completed.
Definition: memory.hpp:54
vcl_size_t element_size< unsigned long >(memory_types)
Definition: memory.hpp:306
void memory_copy(handle_type const &src_buffer, handle_type &dst_buffer, vcl_size_t src_offset, vcl_size_t dst_offset, vcl_size_t bytes_to_copy)
Copies 'bytes_to_copy' bytes from address 'src_buffer + src_offset' on the CUDA device to memory star...
Definition: cuda.hpp:130
This file provides the forward declarations for the main types used within ViennaCL.
void memory_read(mem_handle const &src_buffer, vcl_size_t src_offset, vcl_size_t bytes_to_read, void *ptr, bool async=false)
Reads data from a buffer back to main RAM.
Definition: memory.hpp:261
void memory_read(viennacl::ocl::handle< cl_mem > const &src_buffer, vcl_size_t src_offset, vcl_size_t bytes_to_copy, void *ptr, bool async=false)
Reads data from an OpenCL buffer back to main RAM.
Definition: opencl.hpp:129
void resize(mem_handle const &handle, vcl_size_t num)
Resize including initialization of new memory (cf. std::vector<>)
Definition: util.hpp:138
void memory_write(handle_type &dst_buffer, vcl_size_t dst_offset, vcl_size_t bytes_to_copy, const void *ptr, bool)
Writes data from main RAM identified by 'ptr' to the buffer identified by 'dst_buffer'.
Definition: cpu_ram.hpp:131
Implementation of a OpenCL-like context, which serves as a unification of {OpenMP, CUDA, OpenCL} at the user API.
vcl_size_t element_size(memory_types)
Definition: memory.hpp:299
Represents a generic 'context' similar to an OpenCL context, but is backend-agnostic and thus also su...
Definition: context.hpp:39
vcl_size_t element_size< unsigned int >(memory_types)
Definition: memory.hpp:335
Main namespace in ViennaCL. Holds all the basic types such as vector, matrix, etc. and defines operations upon them.
Definition: cpu_ram.hpp:34
vcl_size_t element_size< long >(memory_types)
Definition: memory.hpp:320
vcl_size_t element_size< int >(memory_types)
Definition: memory.hpp:349
void memory_copy(viennacl::ocl::handle< cl_mem > const &src_buffer, viennacl::ocl::handle< cl_mem > &dst_buffer, vcl_size_t src_offset, vcl_size_t dst_offset, vcl_size_t bytes_to_copy)
Copies 'bytes_to_copy' bytes from address 'src_buffer + src_offset' in the OpenCL context to memory s...
Definition: opencl.hpp:69
Definition: blas3.hpp:36
viennacl::ocl::command_queue & get_queue()
Convenience function for getting the default queue for the currently active device in the active cont...
Definition: backend.hpp:320
Implementations for the OpenCL backend functionality.
Extracts the underlying context from objects.
Implements the multi-memory-domain handle.
std::size_t vcl_size_t
Definition: forwards.h:75
void switch_memory_context(mem_handle &handle, viennacl::context new_ctx)
Switches the active memory domain within a memory handle. Data is copied if the new active domain dif...
Definition: memory.hpp:368
Implementations for the CUDA backend functionality.
void memory_copy(handle_type const &src_buffer, handle_type &dst_buffer, vcl_size_t src_offset, vcl_size_t dst_offset, vcl_size_t bytes_to_copy)
Copies 'bytes_to_copy' bytes from address 'src_buffer + src_offset' to memory starting at address 'ds...
Definition: cpu_ram.hpp:108
Implementations of the OpenCL backend, where all contexts are stored in.
memory_types default_memory_type()
Returns the default memory type for the given configuration.
Definition: mem_handle.hpp:73
viennacl::memory_types memory_type() const
Definition: context.hpp:76
handle_type memory_create(vcl_size_t size_in_bytes, const void *host_ptr=NULL)
Creates an array of the specified size on the CUDA device. If the second argument is provided...
Definition: cuda.hpp:103
void switch_active_handle_id(memory_types new_id)
Switches the currently active handle. If no support for that backend is provided, an exception is thr...
Definition: mem_handle.hpp:121
void memory_copy(mem_handle const &src_buffer, mem_handle &dst_buffer, vcl_size_t src_offset, vcl_size_t dst_offset, vcl_size_t bytes_to_copy)
Copies 'bytes_to_copy' bytes from address 'src_buffer + src_offset' to memory starting at address 'ds...
Definition: memory.hpp:140
viennacl::context context(T const &t)
Returns an ID for the currently active memory domain of an object.
Definition: context.hpp:40
void memory_read(handle_type const &src_buffer, vcl_size_t src_offset, vcl_size_t bytes_to_copy, void *ptr, bool async=false)
Reads data from a CUDA buffer back to main RAM.
Definition: cuda.hpp:183
void set(vcl_size_t index, U value)
Definition: util.hpp:115
Main abstraction class for multiple memory domains. Represents a buffer in either main RAM...
Definition: mem_handle.hpp:89
vcl_size_t raw_size() const
Returns the number of bytes of the currently active buffer.
Definition: mem_handle.hpp:230
void memory_create(mem_handle &handle, vcl_size_t size_in_bytes, viennacl::context const &ctx, const void *host_ptr=NULL)
Creates an array of the specified size. If the second argument is provided, the buffer is initialized...
Definition: memory.hpp:87
Extracts the underlying OpenCL handle from a vector, a matrix, an expression etc. ...
Implementations for the OpenCL backend functionality.
viennacl::backend::mem_handle & handle(T &obj)
Returns the generic memory handle of an object. Non-const version.
Definition: handle.hpp:41
memory_types
Definition: forwards.h:345
vcl_size_t raw_size() const
Definition: util.hpp:111
void memory_shallow_copy(mem_handle const &src_buffer, mem_handle &dst_buffer)
A 'shallow' copy operation from an initialized buffer to an uninitialized buffer. The uninitialized b...
Definition: memory.hpp:177
ram_handle_type & ram_handle()
Returns the handle to a buffer in CPU RAM. NULL is returned if no such buffer has been allocated...
Definition: mem_handle.hpp:99
memory_types get_active_handle_id() const
Returns an ID for the currently active memory buffer. Other memory buffers might contain old or no da...
Definition: mem_handle.hpp:118
Helper functionality for working with different memory domains.
void switch_memory_context(T &obj, viennacl::context new_ctx)
Generic convenience routine for migrating data of an object to a new memory domain.
Definition: memory.hpp:622