CudaMemory.cuh

/*
 * CudaMemory.cu
 *
 *  Created on: Aug 17, 2014
 *      Author: Pietro Incardona
 */
 
#ifndef CUDA_MEMORY_CUH_
#define CUDA_MEMORY_CUH_
 
#define EXCEPT_MC noexcept
 
#include "config.h"
#include "memory.hpp"
#include <iostream>
 
#include "util/cuda_util.hpp"
 
extern size_t TotCudaMemoryAllocated;
 
__device__ inline size_t align_number_device(size_t al, size_t number)
{
    return number + ((number % al) != 0)*(al - number % al);
}
 
static __device__ unsigned char global_cuda_error_array[256];
 
class CudaMemory : public memory
{
    bool is_hm_sync;
    
    size_t sz;
    
    void * dm;
    
    mutable void * hm;
 
    size_t ref_cnt;
    
    void allocate_host(size_t sz) const;
    
    bool copyFromPointer(const void * ptr);
    
public:
    
    bool copyDeviceToDevice(const CudaMemory & m);
 
    virtual bool flush();
    virtual bool allocate(size_t sz);
    virtual void destroy();
    virtual bool copy(const memory & m);
    virtual size_t size() const;
    virtual bool resize(size_t sz);
    virtual void * getPointer();
    
    virtual const void * getPointer() const;
    
    virtual void * getDevicePointer();
 
    virtual void hostToDevice();
 
    virtual void deviceToHost();
 
    virtual void deviceToHost(size_t start, size_t stop);
 
    virtual void hostToDevice(size_t start, size_t top);
 
    void hostToDevice(CudaMemory & mem);
 
    void deviceToHost(CudaMemory & mem);
 
    virtual void fill(unsigned char c);
 
    void isNotSync() {is_hm_sync = false;}
    
    public:
    
    virtual void incRef()
    {
        ref_cnt++;
    }
 
    virtual void decRef()
    {ref_cnt--;}
    
    virtual long int ref()
    {
        return ref_cnt;
    }
 
    bool isInitialized()
    {
        return false;
    }
    
    // Copy the memory (device and host)
    CudaMemory & operator=(const CudaMemory & mem)
    {
        copy(mem);
        return *this;
    }
 
    // Copy the Cuda memory
    CudaMemory(const CudaMemory & mem)
    :CudaMemory()
    {
        allocate(mem.size());
        copy(mem);
    }
 
    CudaMemory(CudaMemory && mem) EXCEPT_MC
    {
        is_hm_sync = mem.is_hm_sync;
        sz = mem.sz;
        dm = mem.dm;
        hm = mem.hm;
        ref_cnt = mem.ref_cnt;
 
        // reset mem
        mem.is_hm_sync = false;
        mem.sz = 0;
        mem.dm = NULL;
        mem.hm = NULL;
        mem.ref_cnt = 0;
    }
    
    CudaMemory():is_hm_sync(true),sz(0),dm(0),hm(0),ref_cnt(0) {};
    
    CudaMemory(size_t sz):is_hm_sync(true),sz(0),dm(0),hm(0),ref_cnt(0)
    {
        allocate(sz);
    };
 
    ~CudaMemory()   
    {
        if(ref_cnt == 0)
            destroy();
        else
            std::cerr << "Error: " << __FILE__ << " " << __LINE__ << " destroying a live object" << "\n"; 
    };
 
    void deviceToDevice(void * ptr, size_t start, size_t stop, size_t offset);
 
    void swap(CudaMemory & mem);
 
    constexpr static bool isDeviceHostSame()
    {
        return false;
    }
 
    void * toKernel()
    {
        return getDevicePointer();
    }
};
 
 
#endif