Commit 05e7d589 authored by David Rohr's avatar David Rohr

fix GPUWORKLOAD parameter

parent 02123bb1
......@@ -305,9 +305,7 @@ int bioem_cuda::deviceInit()
gpumap->sum_RefMap = sum;
gpumap->sumsquare_RefMap = sumsquare;
pProb_device = pProb;
checkCudaErrors(cudaMalloc(&pProb_device.ptr, pProb_device.get_size(RefMap.ntotRefMap, param.nTotGridAngles)));
pProb_device.set_pointers();
checkCudaErrors(cudaMalloc(&pProb_memory, pProb_device.get_size(RefMap.ntotRefMap, param.nTotGridAngles)));
for (int i = 0; i < 2; i++)
{
checkCudaErrors(cudaStreamCreate(&cudaStream[i]));
......@@ -342,7 +340,7 @@ int bioem_cuda::deviceExit()
if (deviceInitialized == 0) return(0);
cudaFree(pProb_device.ptr);
cudaFree(pProb_memory);
cudaFree(sum);
cudaFree(sumsquare);
for (int i = 0; i < 2; i++)
......@@ -383,9 +381,23 @@ int bioem_cuda::deviceExit()
int bioem_cuda::deviceStartRun()
{
maxRef = GPUWorkload >= 100 ? RefMap.ntotRefMap : ((size_t) RefMap.ntotRefMap * (size_t) GPUWorkload / 100);
if (GPUWorkload >= 100)
{
maxRef = RefMap.ntotRefMap;
pProb_host = &pProb;
}
else
{
maxRef = (size_t) RefMap.ntotRefMap * (size_t) GPUWorkload / 100;
pProb_host = new bioem_Probability;
pProb_host->init(maxRef, param.nTotGridAngles, *this);
pProb_host->copyFrom(&pProb, *this);
}
cudaMemcpyAsync(pProb_device.ptr, pProb.ptr, pProb.get_size(RefMap.ntotRefMap, param.nTotGridAngles), cudaMemcpyHostToDevice, cudaStream[0]);
pProb_device = *pProb_host;
pProb_device.ptr = pProb_memory;
pProb_device.set_pointers();
checkCudaErrors(cudaMemcpyAsync(pProb_device.ptr, pProb_host->ptr, pProb_host->get_size(maxRef, param.nTotGridAngles), cudaMemcpyHostToDevice, cudaStream[0]));
if (FFTAlgo)
{
......@@ -420,7 +432,7 @@ int bioem_cuda::deviceStartRun()
int bioem_cuda::deviceFinishRun()
{
if (GPUAsync) cudaStreamSynchronize(cudaStream[0]);
cudaMemcpyAsync(pProb.ptr, pProb_device.ptr, pProb.get_size(RefMap.ntotRefMap, param.nTotGridAngles), cudaMemcpyDeviceToHost, cudaStream[0]);
checkCudaErrors(cudaMemcpyAsync(pProb_host->ptr, pProb_device.ptr, pProb_host->get_size(maxRef, param.nTotGridAngles), cudaMemcpyDeviceToHost, cudaStream[0]));
if (FFTAlgo)
{
......@@ -434,6 +446,12 @@ int bioem_cuda::deviceFinishRun()
if (!GPUDualStream) break;
}
}
cudaThreadSynchronize();
if (GPUWorkload < 100)
{
pProb.copyFrom(pProb_host, *this);
delete[] pProb_host;
}
return(0);
}
......
......@@ -13,6 +13,7 @@
class bioem
{
friend class bioem_RefMap;
friend class bioem_Probability;
public:
bioem();
......
......@@ -34,7 +34,9 @@ protected:
cudaEvent_t cudaFFTEvent[2];
bioem_RefMap_Mod* pRefMap_device_Mod;
bioem_RefMap* gpumap;
bioem_Probability* pProb_host;
bioem_Probability pProb_device;
void* pProb_memory;
myfloat_t* pConvMap_device[2];
mycomplex_t* pRefMapsFFT;
......
......@@ -107,12 +107,13 @@ public:
bioem_Probability_map* ptr_map;
bioem_Probability_angle* ptr_angle;
size_t get_size(size_t maps, size_t angles)
static size_t get_size(size_t maps, size_t angles)
{
return(maps * (angles * sizeof(bioem_Probability_angle) + sizeof(bioem_Probability_map)));
}
void init(size_t maps, size_t angles, bioem& bio);
void copyFrom(bioem_Probability* from, bioem& bio);
void set_pointers()
{
......
......@@ -280,6 +280,20 @@ void bioem_Probability::init(size_t maps, size_t angles, bioem& bio)
set_pointers();
}
void bioem_Probability::copyFrom(bioem_Probability* from, bioem& bio)
{
bioem_Probability_map& pProbMap = getProbMap(0);
bioem_Probability_map& pProbMapFrom = from->getProbMap(0);
memcpy(&pProbMap, &pProbMapFrom, from->nMaps * sizeof(bioem_Probability_map));
for (int iOrient = 0; iOrient < bio.param.nTotGridAngles; iOrient ++)
{
bioem_Probability_angle& pProbAngle = getProbAngle(0, iOrient);
bioem_Probability_angle& pProbAngleFrom = from->getProbAngle(0, iOrient);
memcpy(&pProbAngle, &pProbAngleFrom, from->nMaps * sizeof(bioem_Probability_angle));
}
}
int bioem_RefMap::read_MRC(const char* filename,bioem_param& param)
{
myfloat_t st,st2;
......
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