HologramGenerator_GPU.cpp

Go to the documentation of this file.

#include    "Hologram/HologramGenerator.h"
#include    "graphics/sys.h"
#include    <cuda_runtime.h>
#include    <cufft.h>

static void HandleError(cudaError_t err,
    const char *file,
    int line) {
    if (err != cudaSuccess) {
        printf("%s in %s at line %d\n", cudaGetErrorString(err),
            file, line);
        exit(EXIT_FAILURE);
    }
}
#define HANDLE_ERROR( err ) (HandleError( err, __FILE__, __LINE__ ))


#define HANDLE_NULL( a ) {if (a == NULL) { \
                            printf( "Host memory failed in %s at line %d\n", \
                                    __FILE__, __LINE__ ); \
                            exit( EXIT_FAILURE );}} 

cufftDoubleComplex *u_o_gpu_;
cufftDoubleComplex *u_complex_gpu_;
cufftDoubleComplex *k_temp_d_;

cudaStream_t    stream_;
cudaEvent_t     start, stop;

extern "C"
{
    void cudaFFT(CUstream_st* stream, int nx, int ny, cufftDoubleComplex* in_filed, cufftDoubleComplex* output_field, int direction,  bool bNormailized = false);

    void cudaCropFringe(CUstream_st* stream, int nx, int ny, cufftDoubleComplex* in_field, cufftDoubleComplex* out_field, int cropx1, int cropx2, int cropy1, int cropy2);

    void cudaDepthHoloKernel(CUstream_st* stream, int pnx, int pny, cufftDoubleComplex* u_o_gpu_, unsigned char* img_src_gpu_, unsigned char* dimg_src_gpu_, double* depth_index_gpu_,
        int dtr, double rand_phase_val_a, double rand_phase_val_b, double carrier_phase_delay_a, double carrier_phase_delay_b, int flag_change_depth_quan, unsigned int default_depth_quan);

    void cudaPropagation_AngularSpKernel(CUstream_st* stream_, int pnx, int pny, cufftDoubleComplex* input_d, cufftDoubleComplex* u_complex,
        double ppx, double ppy, double ssx, double ssy, double lambda, double params_k, double propagation_dist);

    void cudaGetFringe(CUstream_st* stream, int pnx, int pny, cufftDoubleComplex* in_field, cufftDoubleComplex* out_field, int sig_locationx, int sig_locationy,
        double ssx, double ssy, double ppx, double ppy, double PI);

    void cudaChangeDepthQuanKernel(CUstream_st* stream_, int pnx, int pny, double* depth_index_gpu, unsigned char* dimg_src_gpu,
        int dtr, double d1, double d2, double params_num_of_depth, double params_far_depthmap, double params_near_depthmap);

}

void HologramGenerator::init_GPU()
{
    const int nx = params_.pn[0];
    const int ny = params_.pn[1];
    const int N = nx * ny;

    if (!stream_)
        cudaStreamCreate(&stream_);
    
    if (img_src_gpu_)   cudaFree(img_src_gpu_);
    HANDLE_ERROR(cudaMalloc((void**)&img_src_gpu_, sizeof(uchar1)*N));

    if (dimg_src_gpu_)  cudaFree(dimg_src_gpu_);
    HANDLE_ERROR(cudaMalloc((void**)&dimg_src_gpu_, sizeof(uchar1)*N));

    if (depth_index_gpu_) cudaFree(depth_index_gpu_);
    if (FLAG_CHANGE_DEPTH_QUANTIZATION == 1)
        HANDLE_ERROR(cudaMalloc((void**)&depth_index_gpu_, sizeof(double)*N));
    
    if (u_o_gpu_)   cudaFree(u_o_gpu_);
    if (u_complex_gpu_) cudaFree(u_complex_gpu_);

    HANDLE_ERROR(cudaMalloc((void**)&u_o_gpu_, sizeof(cufftDoubleComplex)*N));
    HANDLE_ERROR(cudaMalloc((void**)&u_complex_gpu_, sizeof(cufftDoubleComplex)*N));

    if (k_temp_d_)  cudaFree(k_temp_d_);
    HANDLE_ERROR(cudaMalloc((void**)&k_temp_d_, sizeof(cufftDoubleComplex)*N));

}

bool HologramGenerator::prepare_inputdata_GPU(uchar* imgptr, uchar* dimgptr)
{
    const int nx = params_.pn[0];
    const int ny = params_.pn[1];
    const int N = nx * ny;
    
    HANDLE_ERROR(cudaMemcpyAsync(img_src_gpu_, imgptr, sizeof(uchar1)*N, cudaMemcpyHostToDevice), stream_);
    HANDLE_ERROR(cudaMemcpyAsync(dimg_src_gpu_, dimgptr, sizeof(uchar1)*N, cudaMemcpyHostToDevice), stream_);
    
    return true;
}

void HologramGenerator::change_depth_quan_GPU()
{
    int pnx = params_.pn[0];
    int pny = params_.pn[1];

    double temp_depth, d1, d2;
    int tdepth;

    HANDLE_ERROR(cudaMemsetAsync(depth_index_gpu_, 0, sizeof(double)*pnx*pny, stream_));

    for (int dtr = 0; dtr < params_.num_of_depth; dtr++)
    {
        temp_depth = dlevel_[dtr];
        d1 = temp_depth - dstep_ / 2.0;
        d2 = temp_depth + dstep_ / 2.0;

        cudaChangeDepthQuanKernel(stream_, pnx, pny, depth_index_gpu_, dimg_src_gpu_, 
            dtr, d1, d2, params_.num_of_depth, params_.far_depthmap, params_.near_depthmap);
    }

}

void HologramGenerator::Calc_Holo_GPU(int frame)
{
    cudaEvent_t start, stop;

    cudaEventCreate(&start);
    cudaEventCreate(&stop);

    if (!stream_)
        cudaStreamCreate(&stream_);

    cudaEventRecord(start, stream_);

    int pnx = params_.pn[0];
    int pny = params_.pn[1];
    int N = pnx*pny;

    HANDLE_ERROR(cudaMemsetAsync(u_complex_gpu_, 0, sizeof(cufftDoubleComplex)*N, stream_));
    int depth_sz = params_.render_depth.size();

    for (int p = 0; p < depth_sz; ++p)
    {
        Complex rand_phase_val;
        get_rand_phase_value(rand_phase_val);

        int dtr = params_.render_depth[p];
        double temp_depth = dlevel_transform_[dtr - 1];
        Complex carrier_phase_delay(0, params_.k* temp_depth);
        exponent_complex(&carrier_phase_delay);

        HANDLE_ERROR(cudaMemsetAsync(u_o_gpu_, 0, sizeof(cufftDoubleComplex)*N, stream_));

        cudaDepthHoloKernel(stream_, pnx, pny, u_o_gpu_, img_src_gpu_, dimg_src_gpu_, depth_index_gpu_, 
            dtr, rand_phase_val.a, rand_phase_val.b, carrier_phase_delay.a, carrier_phase_delay.b, FLAG_CHANGE_DEPTH_QUANTIZATION, DEFAULT_DEPTH_QUANTIZATION);

        if (Propagation_Method_ == 0)
        {
            HANDLE_ERROR(cudaMemsetAsync(k_temp_d_, 0, sizeof(cufftDoubleComplex)*N, stream_));
            cudaFFT(stream_, pnx, pny, u_o_gpu_, k_temp_d_, -1);

            Propagation_AngularSpectrum_GPU(u_o_gpu_, -temp_depth);
        }

        LOG("Frame#: %d, Depth: %d of %d, z = %f mm\n", frame, dtr, params_.num_of_depth, -temp_depth * 1000);

    }

    cudaEventRecord(stop, stream_);
    cudaEventSynchronize(stop);

    float elapsedTime = 0.0f;
    cudaEventElapsedTime(&elapsedTime, start, stop);
    LOG("GPU Time= %f ms. \n", elapsedTime);

    //writeIntensity_gray8_real_bmp("final_fr", pnx, pny, U_complex_);


}

void HologramGenerator::Propagation_AngularSpectrum_GPU(cufftDoubleComplex* input_u, double propagation_dist)
{
    int pnx = params_.pn[0];
    int pny = params_.pn[1];
    int N = pnx* pny;
    double ppx = params_.pp[0];
    double ppy = params_.pp[1];
    double ssx = params_.ss[0];
    double ssy = params_.ss[1];
    double lambda = params_.lambda;

    cudaPropagation_AngularSpKernel(stream_, pnx, pny, k_temp_d_, u_complex_gpu_, 
        ppx, ppy, ssx, ssy, lambda, params_.k, propagation_dist);
        
}

void HologramGenerator::encoding_GPU(int cropx1, int cropx2, int cropy1, int cropy2, ivec2 sig_location)
{
    int pnx = params_.pn[0];
    int pny = params_.pn[1];
    double ppx = params_.pp[0];
    double ppy = params_.pp[1];
    double ssx = params_.ss[0];
    double ssy = params_.ss[1];

    HANDLE_ERROR(cudaMemsetAsync(k_temp_d_, 0, sizeof(cufftDoubleComplex)*pnx*pny, stream_));
    cudaCropFringe(stream_, pnx, pny, u_complex_gpu_, k_temp_d_, cropx1, cropx2, cropy1, cropy2);

    HANDLE_ERROR(cudaMemsetAsync(u_complex_gpu_, 0, sizeof(cufftDoubleComplex)*pnx*pny, stream_));
    cudaFFT(stream_, pnx, pny, k_temp_d_, u_complex_gpu_, 1, true);

    HANDLE_ERROR(cudaMemsetAsync(k_temp_d_, 0, sizeof(cufftDoubleComplex)*pnx*pny, stream_));
    cudaGetFringe(stream_, pnx, pny, u_complex_gpu_, k_temp_d_, sig_location[0], sig_location[1], ssx, ssy, ppx, ppy, PI);

    cufftDoubleComplex* sample_fd = (cufftDoubleComplex*)malloc(sizeof(cufftDoubleComplex)*pnx*pny);
    memset(sample_fd, 0.0, sizeof(cufftDoubleComplex)*pnx*pny);

    HANDLE_ERROR(cudaMemcpyAsync(sample_fd, k_temp_d_, sizeof(cufftDoubleComplex)*pnx*pny, cudaMemcpyDeviceToHost), stream_);
    memset(u255_fringe_, 0.0, sizeof(double)*pnx*pny);

    for (int i = 0; i < pnx*pny; ++i)
    {
        u255_fringe_[i] = sample_fd[i].x;
    }

    free(sample_fd);
}

/*
void HologramGenerator::writeImage_fromGPU(QString imgname, int pnx, int pny, cufftDoubleComplex* gpu_data)
{
    cufftDoubleComplex* cpu_data = (cufftDoubleComplex*)malloc(sizeof(cufftDoubleComplex)*pnx*pny);
    memset(cpu_data, 0.0, sizeof(cufftDoubleComplex)*pnx*pny);

    double* data = (double*)malloc(sizeof(double)*pnx*pny);

    HANDLE_ERROR(cudaMemcpyAsync(cpu_data, gpu_data, sizeof(cufftDoubleComplex)*pnx*pny, cudaMemcpyDeviceToHost), stream_);
    for (int i = 0; i < pnx*pny; ++i)
    {
        data[i] = cpu_data[i].x;
    }

    writeIntensity_gray8_bmp(imgname.toStdString().c_str(),pnx, pny, data );

    free(cpu_data);
    free(data);

}*/