Commit 25d2e7dc authored by David Rohr's avatar David Rohr

finish first mpi implementation

parent 5fd06c57
This diff is collapsed.
......@@ -30,6 +30,20 @@ __device__ static inline void update_prob(const myfloat_t logpro, const int iRef
{
pProbMap.Total = pProbMap.Total * exp(-logpro + pProbMap.Constoadd);
pProbMap.Constoadd = logpro;
// ********** Getting parameters that maximize the probability ***********
if (GPUAlgo == 2)
{
bufint[0] = 1;
buf3[1] = logpro;
}
else
{
pProbMap.max.max_prob_cent_x = cent_x;
pProbMap.max.max_prob_cent_y = cent_y;
}
pProbMap.max.max_prob_orient = iOrient;
pProbMap.max.max_prob_conv = iConv;
}
if (GPUAlgo != 2) pProbMap.Total += exp(logpro - pProbMap.Constoadd);
......@@ -46,25 +60,6 @@ __device__ static inline void update_prob(const myfloat_t logpro, const int iRef
if (GPUAlgo != 2) pProbAngle.forAngles += exp(logpro - pProbAngle.ConstAngle);
}
// ********** Getting parameters that maximize the probability ***********
if(pProbMap.max_prob < logpro)
{
pProbMap.max_prob = logpro;
if (GPUAlgo == 2)
{
bufint[0] = 1;
buf3[1] = logpro;
}
else
{
pProbMap.max_prob_cent_x = cent_x;
pProbMap.max_prob_cent_y = cent_y;
}
pProbMap.max_prob_orient = iOrient;
pProbMap.max_prob_conv = iConv;
}
}
__device__ static inline myfloat_t calc_logpro(const bioem_param_device& param, const myfloat_t sum, const myfloat_t sumsquare, const myfloat_t crossproMapConv, const myfloat_t sumref, const myfloat_t sumsquareref)
......@@ -97,6 +92,12 @@ __device__ static inline void calProb(int iRefMap, int iOrient, int iConv, myflo
{
pProbMap.Total = pProbMap.Total * exp(-logpro + pProbMap.Constoadd);
pProbMap.Constoadd = logpro;
// ********** Getting parameters that maximize the probability ***********
pProbMap.max.max_prob_cent_x = disx;
pProbMap.max.max_prob_cent_y = disy;
pProbMap.max.max_prob_orient = iOrient;
pProbMap.max.max_prob_conv = iConv;
}
pProbMap.Total += exp(logpro - pProbMap.Constoadd);
......@@ -111,15 +112,6 @@ __device__ static inline void calProb(int iRefMap, int iOrient, int iConv, myflo
}
pProbAngle.forAngles += exp(logpro - pProbAngle.ConstAngle);
}
if(pProbMap.max_prob < logpro)
{
pProbMap.max_prob = logpro;
pProbMap.max_prob_cent_x = disx;
pProbMap.max_prob_cent_y = disy;
pProbMap.max_prob_orient = iOrient;
pProbMap.max_prob_conv = iConv;
}
}
__device__ static inline void doRefMapFFT(const int iRefMap, const int iOrient, const int iConv, const myfloat_t* lCC, const myfloat_t sumC, const myfloat_t sumsquareC, bioem_Probability& pProb, const bioem_param_device& param, const bioem_RefMap& RefMap)
......@@ -305,8 +297,8 @@ __device__ static inline void compareRefMap(const int iRefMap, const int iOrient
if (bufint[0] == 1 && buf3[1] == logpro && iRefMap < RefMap.ntotRefMap && atomicAdd(&bufint[0], 1) == 1)
{
pProbMap.max_prob_cent_x = cent_x;
pProbMap.max_prob_cent_y = cent_y;
pProbMap.max.max_prob_cent_x = cent_x;
pProbMap.max.max_prob_cent_y = cent_y;
}
__syncthreads();
......
......@@ -336,6 +336,11 @@ int bioem_cuda::selectCudaDevice()
printf("textureAlignment = %lld", (unsigned long long int) deviceProp.textureAlignment);
}
if (DebugOutput >= 1)
{
printf("BioEM for CUDA initialized, %d GPUs found, using GPU %d\n", count, bestDevice);
}
return(0);
}
......
......@@ -20,6 +20,7 @@ typedef float myfloat_t;
#define MY_CUFFT_C2R CUFFT_C2R
#define mycufftExecC2R cufftExecC2R
#define mycuComplex_t cuComplex
#define MY_MPI_FLOAT MPI_FLOAT
#else
typedef double myfloat_t;
#define myfftw_malloc fftw_malloc
......@@ -37,6 +38,7 @@ typedef double myfloat_t;
#define mycufftExecC2R cufftExecZ2D
#define mycuComplex_t cuDoubleComplex
#define MY_CUFFT_C2R CUFFT_Z2D
#define MY_MPI_FLOAT MPI_DOUBLE
#endif
typedef myfloat_t mycomplex_t[2];
......
......@@ -5,6 +5,7 @@
#include "param.h"
#include <complex>
#include <math.h>
#include <boost/concept_check.hpp>
class bioem_param;
class bioem;
......@@ -85,8 +86,12 @@ class bioem_Probability_map
public:
myfloat_t Total;
myfloat_t Constoadd;
myfloat_t max_prob;
int max_prob_cent_x, max_prob_cent_y, max_prob_orient, max_prob_conv;
class bioem_Probability_map_max
{
public:
int max_prob_cent_x, max_prob_cent_y, max_prob_orient, max_prob_conv;
} max;
};
class bioem_Probability_angle
......
......@@ -70,18 +70,18 @@ int main(int argc, char* argv[])
}
// ************ Configuration and Pre-calculating necessary objects *****************
printf("Configuring\n");
if (mpi_rank == 0) printf("Configuring\n");
if (bio->configure(argc, argv) == 0)
{
// ******************************* Run BioEM routine ******************************
printf("Running\n");
if (mpi_rank == 0) printf("Running\n");
timer.Start();
bio->run();
timer.Stop();
// ************************************ End **********************************
printf ("The code ran for %f seconds.\n", timer.GetElapsedTime());
printf ("The code ran for %f seconds (rank %d).\n", timer.GetElapsedTime(), mpi_rank);
bio->cleanup();
}
delete bio;
......
......@@ -273,7 +273,7 @@ int bioem_param::readParameters(const char* fileinput)
void bioem_param::PrepareFFTs()
{
cout << "Preparing FFTs\n";
if (mpi_rank) cout << "Preparing FFTs\n";
releaseFFTPlans();
mycomplex_t *tmp_map, *tmp_map2;
tmp_map = (mycomplex_t *) myfftw_malloc(sizeof(mycomplex_t) * param_device.NumberPixels * param_device.NumberPixels);
......
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