Commit e9b15195 authored by David Rohr's avatar David Rohr
Browse files

Fix memory allocation for CC Probability

parent 62cf7c72
...@@ -270,7 +270,7 @@ int bioem::configure(int ac, char* av[]) ...@@ -270,7 +270,7 @@ int bioem::configure(int ac, char* av[])
printf("Time Precalculate %f\n", timer.GetCurrentElapsedTime()); printf("Time Precalculate %f\n", timer.GetCurrentElapsedTime());
timer.ResetStart(); timer.ResetStart();
} }
pProb.init(RefMap.ntotRefMap, param.nTotGridAngles, *this); pProb.init(RefMap.ntotRefMap, param.nTotGridAngles, param.nTotCC, *this);
if (DebugOutput >= 2 && mpi_rank == 0) if (DebugOutput >= 2 && mpi_rank == 0)
{ {
......
...@@ -397,8 +397,7 @@ int bioem_cuda::deviceInit() ...@@ -397,8 +397,7 @@ int bioem_cuda::deviceInit()
gpumap->sum_RefMap = sum; gpumap->sum_RefMap = sum;
gpumap->sumsquare_RefMap = sumsquare; gpumap->sumsquare_RefMap = sumsquare;
checkCudaErrors(cudaMalloc(&pProb_memory, pProb_device.get_size(RefMap.ntotRefMap, param.nTotGridAngles, param.param_device.writeAngles))); checkCudaErrors(cudaMalloc(&pProb_memory, pProb_device.get_size(RefMap.ntotRefMap, param.nTotGridAngles, param.nTotCC, param.param_device.writeAngles, param.param_device.writeCC)));
checkCudaErrors(cudaMalloc(&pProb_memory, pProb_device.get_sizeCC(RefMap.ntotRefMap, param.nTotCC, param.param_device.writeCC)));
for (int i = 0; i < 2; i++) for (int i = 0; i < 2; i++)
{ {
...@@ -484,15 +483,14 @@ int bioem_cuda::deviceStartRun() ...@@ -484,15 +483,14 @@ int bioem_cuda::deviceStartRun()
{ {
maxRef = (size_t) RefMap.ntotRefMap * (size_t) GPUWorkload / 100; maxRef = (size_t) RefMap.ntotRefMap * (size_t) GPUWorkload / 100;
pProb_host = new bioem_Probability; pProb_host = new bioem_Probability;
pProb_host->init(maxRef, param.nTotGridAngles, *this); pProb_host->init(maxRef, param.nTotGridAngles, param.nTotCC, *this);
pProb_host->initCC(maxRef, param.nTotCC, *this);
pProb_host->copyFrom(&pProb, *this); pProb_host->copyFrom(&pProb, *this);
} }
pProb_device = *pProb_host; pProb_device = *pProb_host;
pProb_device.ptr = pProb_memory; pProb_device.ptr = pProb_memory;
pProb_device.set_pointers(); pProb_device.set_pointers();
checkCudaErrors(cudaMemcpyAsync(pProb_device.ptr, pProb_host->ptr, pProb_host->get_size(maxRef, param.nTotGridAngles, param.param_device.writeAngles), cudaMemcpyHostToDevice, cudaStream[0])); checkCudaErrors(cudaMemcpyAsync(pProb_device.ptr, pProb_host->ptr, pProb_host->get_size(maxRef, param.nTotGridAngles, param.nTotCC, param.param_device.writeAngles, param.param_device.writeCC), cudaMemcpyHostToDevice, cudaStream[0]));
if (FFTAlgo) if (FFTAlgo)
{ {
...@@ -527,7 +525,7 @@ int bioem_cuda::deviceStartRun() ...@@ -527,7 +525,7 @@ int bioem_cuda::deviceStartRun()
int bioem_cuda::deviceFinishRun() int bioem_cuda::deviceFinishRun()
{ {
if (GPUAsync) cudaStreamSynchronize(cudaStream[0]); if (GPUAsync) cudaStreamSynchronize(cudaStream[0]);
checkCudaErrors(cudaMemcpyAsync(pProb_host->ptr, pProb_device.ptr, pProb_host->get_size(maxRef, param.nTotGridAngles, param.param_device.writeAngles), cudaMemcpyDeviceToHost, cudaStream[0])); checkCudaErrors(cudaMemcpyAsync(pProb_host->ptr, pProb_device.ptr, pProb_host->get_size(maxRef, param.nTotGridAngles, param.nTotCC, param.param_device.writeAngles, param.param_device.writeCC), cudaMemcpyDeviceToHost, cudaStream[0]));
if (FFTAlgo) if (FFTAlgo)
{ {
......
...@@ -91,7 +91,7 @@ public: ...@@ -91,7 +91,7 @@ public:
{ {
public: public:
int max_prob_cent_x, max_prob_cent_y, max_prob_orient, max_prob_conv; int max_prob_cent_x, max_prob_cent_y, max_prob_orient, max_prob_conv;
myfloat_t max_prob_norm,max_prob_mu; myfloat_t max_prob_norm,max_prob_mu;
} max; } max;
}; };
...@@ -105,7 +105,7 @@ public: ...@@ -105,7 +105,7 @@ public:
class bioem_Probability_cc class bioem_Probability_cc
{ {
public: public:
myfloat_t forCC; myfloat_t forCC;
}; };
class bioem_Probability class bioem_Probability
...@@ -113,39 +113,33 @@ class bioem_Probability ...@@ -113,39 +113,33 @@ class bioem_Probability
public: public:
int nMaps; int nMaps;
int nAngles; int nAngles;
int nCC; int nCC;
__device__ __host__ bioem_Probability_map& getProbMap(int map) {return(ptr_map[map]);} __device__ __host__ bioem_Probability_map& getProbMap(int map) {return(ptr_map[map]);}
__device__ __host__ bioem_Probability_angle& getProbAngle(int map, int angle) {return(ptr_angle[angle * nMaps + map]);} __device__ __host__ bioem_Probability_angle& getProbAngle(int map, int angle) {return(ptr_angle[angle * nMaps + map]);}
__device__ __host__ bioem_Probability_cc& getProbCC(int map, int cc) {return(ptr_cc[cc * nMaps + map]);} __device__ __host__ bioem_Probability_cc& getProbCC(int map, int cc) {return(ptr_cc[cc * nMaps + map]);}
void* ptr; void* ptr;
bioem_Probability_map* ptr_map; bioem_Probability_map* ptr_map;
bioem_Probability_angle* ptr_angle; bioem_Probability_angle* ptr_angle;
bioem_Probability_cc* ptr_cc; bioem_Probability_cc* ptr_cc;
static size_t get_size(size_t maps, size_t angles, bool writeAngles) static size_t get_size(size_t maps, size_t angles, size_t cc, bool writeAngles, bool writeCC)
{ {
size_t size = sizeof(bioem_Probability_map); size_t size = sizeof(bioem_Probability_map);
if (writeAngles) size += angles * sizeof(bioem_Probability_angle); if (writeAngles) size += angles * sizeof(bioem_Probability_angle);
if (writeCC) size += cc * sizeof(bioem_Probability_cc);
return(maps * size); return(maps * size);
} }
static size_t get_sizeCC(size_t maps, size_t cc, bool writeCC)
{ void init(size_t maps, size_t angles, size_t cc, bioem& bio);
size_t size = sizeof(bioem_Probability_map);
if (writeCC) size += cc * sizeof(bioem_Probability_cc) ;
return(maps * size);
}
void init(size_t maps, size_t angles, bioem& bio);
void initCC(size_t maps, size_t cc, bioem& bio);
void copyFrom(bioem_Probability* from, bioem& bio); void copyFrom(bioem_Probability* from, bioem& bio);
void set_pointers() void set_pointers()
{ {
ptr_map = (bioem_Probability_map*) ptr; ptr_map = (bioem_Probability_map*) ptr;
ptr_angle = (bioem_Probability_angle*) (&ptr_map[nMaps]); ptr_angle = (bioem_Probability_angle*) (&ptr_map[nMaps]);
ptr_cc = (bioem_Probability_cc*) (&ptr_map[nMaps]); ptr_cc = (bioem_Probability_cc*) (&ptr_angle[nMaps * nAngles]);
} }
}; };
......
...@@ -280,23 +280,15 @@ int bioem_RefMap::precalculate(bioem_param& param, bioem& bio) ...@@ -280,23 +280,15 @@ int bioem_RefMap::precalculate(bioem_param& param, bioem& bio)
return(0); return(0);
} }
void bioem_Probability::init(size_t maps, size_t angles, bioem& bio) void bioem_Probability::init(size_t maps, size_t angles, size_t cc, bioem& bio)
{ {
nMaps = maps; nMaps = maps;
nAngles = angles; nAngles = angles;
ptr = bio.malloc_device_host(get_size(maps, angles, bio.param.param_device.writeAngles)); nCC = cc;
ptr = bio.malloc_device_host(get_size(maps, angles, cc, bio.param.param_device.writeAngles, bio.param.param_device.writeCC));
set_pointers(); set_pointers();
} }
void bioem_Probability::initCC(size_t maps, size_t cc, bioem& bio)
{
nMaps = maps;
nCC = cc;
ptr = bio.malloc_device_host(get_sizeCC(maps, cc, bio.param.param_device.writeCC));
set_pointers();
}
void bioem_Probability::copyFrom(bioem_Probability* from, bioem& bio) void bioem_Probability::copyFrom(bioem_Probability* from, bioem& bio)
{ {
bioem_Probability_map& pProbMap = getProbMap(0); bioem_Probability_map& pProbMap = getProbMap(0);
...@@ -305,13 +297,23 @@ void bioem_Probability::copyFrom(bioem_Probability* from, bioem& bio) ...@@ -305,13 +297,23 @@ void bioem_Probability::copyFrom(bioem_Probability* from, bioem& bio)
if (bio.param.param_device.writeAngles) if (bio.param.param_device.writeAngles)
{ {
for (int iOrient = 0; iOrient < bio.param.nTotGridAngles; iOrient ++) for (int iOrient = 0; iOrient < nAngles; iOrient ++)
{ {
bioem_Probability_angle& pProbAngle = getProbAngle(0, iOrient); bioem_Probability_angle& pProbAngle = getProbAngle(0, iOrient);
bioem_Probability_angle& pProbAngleFrom = from->getProbAngle(0, iOrient); bioem_Probability_angle& pProbAngleFrom = from->getProbAngle(0, iOrient);
memcpy(&pProbAngle, &pProbAngleFrom, from->nMaps * sizeof(bioem_Probability_angle)); memcpy(&pProbAngle, &pProbAngleFrom, from->nMaps * sizeof(bioem_Probability_angle));
} }
} }
if (bio.param.param_device.writeCC)
{
for (int iCC = 0; iCC < nCC; iCC ++)
{
bioem_Probability_cc& pProbCC = getProbCC(0, iCC);
bioem_Probability_cc& pProbCCFrom = from->getProbCC(0, iCC);
memcpy(&pProbCC, &pProbCCFrom, from->nMaps * sizeof(bioem_Probability_cc));
}
}
} }
int bioem_RefMap::read_MRC(const char* filename,bioem_param& param) int bioem_RefMap::read_MRC(const char* filename,bioem_param& param)
......
Markdown is supported
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment