Memory Management Basics

Device pointers cannot be dereferenced on the host (and vice versa) unless using unified/mapped memory.

6.1 Core API

cudaMalloc(&ptr, bytes);                               // allocate device (global) memory
cudaFree(ptr);
cudaMemcpy(dst, src, bytes, kind);                     // kind: cudaMemcpyHostToDevice, DeviceToHost,
                                                       //       DeviceToDevice, cudaMemcpyDefault
cudaMemset(ptr, 0, bytes);                             // set bytes
cudaMallocHost(&h_ptr, bytes); cudaFreeHost(h_ptr);    // pinned (page-locked) host memory (§18)
cudaMallocManaged(&ptr, bytes);                        // unified memory (§19)
cudaMallocPitch(&ptr, &pitch, widthBytes, height);     // padded 2-D allocation
cudaMemcpy2D(dst, dpitch, src, spitch, widthBytes, height, kind);
cudaMemcpyAsync(dst, src, bytes, kind, stream);        // asynchronous (§17)
cudaMemGetInfo(&freeBytes, &totalBytes);

6.2 Host vs device pointers

Device pointers cannot be dereferenced on the host (and vice versa) unless using unified/mapped memory. Keep naming conventions: h_ for host, d_ for device.

6.3 RAII wrapper (C++)

Or use thrust::device_vector<T> (§31).

template <typename T>
struct DeviceBuffer {
    T* ptr = nullptr;
    size_t n = 0;
    explicit DeviceBuffer(size_t count) : n(count) { cudaMalloc(&ptr, count * sizeof(T)); }
    ~DeviceBuffer() { cudaFree(ptr); }
    DeviceBuffer(const DeviceBuffer&) = delete;
    DeviceBuffer& operator=(const DeviceBuffer&) = delete;
    void upload(const T* h)   { cudaMemcpy(ptr, h, n * sizeof(T), cudaMemcpyHostToDevice); }
    void download(T* h) const { cudaMemcpy(h, ptr, n * sizeof(T), cudaMemcpyDeviceToHost); }
};