Why is CUDA pinned memory so fast?

by

CUDA Driver checks, if the memory range is locked or not and then it will use a different codepath. Locked memory is stored in the physical memory (RAM), so device can fetch it w/o help from CPU (DMA, aka Async copy; device only need list of physical pages). Not-locked memory can generate a page fault on access, and it is stored not only in memory (e.g. it can be in swap), so driver need to access every page of non-locked memory, copy it into pinned buffer and pass it to DMA (Syncronious, page-by-page copy).

As described here http://forums.nvidia.com/index.php?showtopic=164661

host memory used by the asynchronous mem copy call needs to be page locked through cudaMallocHost or cudaHostAlloc.

I can also recommend to check cudaMemcpyAsync and cudaHostAlloc manuals at developer.download.nvidia.com. HostAlloc says that cuda driver can detect pinned memory:

The driver tracks the virtual memory ranges allocated with this(cudaHostAlloc) function and automatically accelerates calls to functions such as cudaMemcpy().

More Related Contents:

  1. Significance of -pthread flag when compiling
  2. Linux API to list running processes?
  3. What are the GCC default include directories?
  4. VS Code will not build c++ programs with multiple .ccp source files
  5. High resolution timer with C++ and Linux?
  6. How do I start a new CUDA project in Visual Studio 2008?
  7. How do I start a CUDA app in Visual Studio 2010?
  8. Are child processes created with fork() automatically killed when the parent is killed?
  9. Linking a shared library with another shared lib in linux
  10. Issuing system commands in Linux from C, C++
  11. will casting around sockaddr_storage and sockaddr_in break strict aliasing
  12. Spinlock versus Semaphore
  13. CUDA and Classes
  14. Proper way to create unique_ptr that holds an allocated array
  15. How to print pthread_t
  16. Netbeans 7.2 shows “Unable to resolve identifier” , although build is successful
  17. How to pause a pthread ANY TIME I want?
  18. How to set breakpoints on future shared libraries with a command flag
  19. How to enable core dump in my Linux C++ program [duplicate]
  20. Compiling Cuda code in Qt Creator on Windows
  21. Simple Linux Signal Handling
  22. Is there a limit of stack size of a process in linux
  23. How to create a temporary directory in C++?
  24. How performing multiple matrix multiplications in CUDA?
  25. /usr/lib/x86_64-linux-gnu/libstdc++.so.6: version CXXABI_1.3.8′ not found
  26. Clearing terminal in Linux with C++ code
  27. Install gcc on linux with no root privilege
  28. Linux c++ error: undefined reference to ‘dlopen’
  29. “to_string” isn’t a member of “std”?
  30. Exception running boost asio ssl example

Leave a Comment