Unified Memory For CUDA Rookies
", introduced the fundamentals of CUDA programming by showing how to write a simple program that allotted two arrays of numbers in memory accessible to the GPU and then added them collectively on the GPU. To do that, I introduced you to Unified Memory, which makes it very straightforward to allocate and access information that may be used by code operating on any processor in the system, CPU or GPU. I finished that publish with just a few simple "exercises", certainly one of which inspired you to run on a latest Pascal-based GPU to see what happens. I hoped that readers would try it and comment on the outcomes, and some of you did! I recommended this for 2 causes. First, as a result of Pascal GPUs such because the NVIDIA Titan X and the NVIDIA Tesla P100 are the first GPUs to include the Page Migration Engine, which is hardware help for Unified Memory page faulting and migration.
The second motive is that it offers an ideal alternative to learn extra about Unified Memory. Quick GPU, Quick Memory… Right! But let’s see. First, I’ll reprint the results of operating on two NVIDIA Kepler GPUs (one in my laptop and one in a server). Now let’s strive working on a really fast Tesla P100 accelerator, primarily based on the Pascal GP100 GPU. Hmmmm, that’s underneath 6 GB/s: slower than operating on my laptop’s Kepler-based GeForce GPU. Don’t be discouraged, although; we can repair this. To understand how, I’ll need to inform you a bit more about Unified Memory. What's Unified Memory? Unified Memory is a single memory deal with area accessible from any processor in a system (see Figure 1). This hardware/software program know-how permits applications to allocate information that can be learn or written from code working on both CPUs or GPUs. Allocating Unified Memory is so simple as replacing calls to malloc() or new with calls to cudaMallocManaged(), an allocation operate that returns a pointer accessible from any processor (ptr in the next).
When code operating on a CPU or GPU accesses data allocated this way (often known as CUDA managed knowledge), the CUDA system software and/or the hardware takes care of migrating memory pages to the memory of the accessing processor. The vital level right here is that the Pascal GPU structure is the first with hardware help for virtual memory web page faulting and migration, by way of its Page Migration Engine. Older GPUs primarily based on the Kepler and Maxwell architectures also support a more restricted type of Unified Memory. What Occurs on Kepler Once i name cudaMallocManaged()? On techniques with pre-Pascal GPUs just like the Tesla K80, calling cudaMallocManaged() allocates size bytes of managed memory on the GPU device that's lively when the call is made1. Internally, the driver also sets up web page table entries for all pages covered by the allocation, in order that the system knows that the pages are resident on that GPU. So, in our instance, working on a Tesla K80 GPU (Kepler architecture), x and y are both initially absolutely resident in GPU memory.
Then within the loop starting on line 6, the CPU steps by means of each arrays, initializing their components to 1.0f and 2.0f, respectively. Because the pages are initially resident in system Memory Wave Protocol, a page fault occurs on the CPU for each array web page to which it writes, and the GPU driver migrates the page from machine memory to CPU memory. After the loop, all pages of the two arrays are resident in CPU memory. After initializing the info on the CPU, the program launches the add() kernel so as to add the weather of x to the weather of y. On pre-Pascal GPUs, upon launching a kernel, the CUDA runtime should migrate all pages beforehand migrated to host memory or to another GPU back to the machine memory of the device operating the kernel2. Since these older GPUs can’t page fault, all data must be resident on the GPU just in case the kernel accesses it (even when it won’t).
Memory Wave
Memory Wave
Memory Wave Protocol