...
 
Commits (48)
......@@ -25,6 +25,7 @@
*.log
*.sql
*.sqlite
*.xml
# OS generated files #
######################
......@@ -74,4 +75,35 @@ install_dir
**/*.vtk
!**/test_data/*.vtk
*.csv
*.h5
*.json
*.html
cl_part_norm_hilbert_dev_2_ref
cl_part_norm_hilbert_dev_3_ref
cl_part_norm_hilbert_mean_2_ref
cl_part_norm_hilbert_mean_3_ref
cl_part_reorder_hilbert_dev_2_ref
cl_part_reorder_hilbert_dev_3_ref
cl_part_reorder_hilbert_mean_2_ref
cl_part_reorder_hilbert_mean_3_ref
cuda_include
cuda_lib
cuda_options
cuda_suppress_wr
error_code
Makefile
grid_dist_id.h51
grid_dist_id.h52
grid_dist_id.h53
grid_iterator_dev_3_ref
grid_iterator_mean_3_ref
grid_m2p_dev_3_ref
grid_m2p_mean_3_ref
mpi_include
mpi_libs
out
perf.data
perf.data.old
projectId.sh
/build
/Default/
......@@ -30,6 +30,8 @@ set(SuiteSparse_ROOT ${SUITESPARSE_ROOT})
set (CMAKE_CXX_STANDARD 11)
set (CMAKE_CUDA_STANDARD 11)
set(CMAKE_CXX_FLAGS_RELEASE "-O3")
set(CMAKE_CXX_FLAGS_RELWITHDEBINFO "-g -O3")
if (OPENBLAS_ROOT)
set(ENV{LD_LIBRARY_PATH} "$ENV{LD_LIBRARY_PATH}:${OPENBLAS_ROOT}/lib")
......@@ -46,18 +48,18 @@ if(ENABLE_GPU)
if (CUDA_VERSION_MAJOR EQUAL 9 AND CUDA_VERSION_MINOR EQUAL 2)
message("CUDA is compatible")
set(WARNING_SUPPRESSION_AND_OPTION_NVCC -Xcudafe "--display_error_number --diag_suppress=611 --diag_suppress=2885 --diag_suppress=2886 --diag_suppress=2887 --diag_suppress=2888 --diag_suppress=186 --diag_suppress=111" --expt-extended-lambda)
FILE(WRITE cuda_options " -Xcudafe \"--display_error_number --diag_suppress=611 --diag_suppress=2885 --diag_suppress=2886 --diag_suppress=2887 --diag_suppress=2888 --diag_suppress=186 --diag_suppress=111\" --expt-extended-lambda ")
elseif ( CUDA_VERSION_MAJOR EQUAL 10 AND CUDA_VERSION_MINOR EQUAL 1 )
message("CUDA is compatible")
set(WARNING_SUPPRESSION_AND_OPTION_NVCC -Xcudafe "--display_error_number --diag_suppress=2915 --diag_suppress=2912 --diag_suppress=2913 --diag_suppress=111 --diag_suppress=186 --diag_suppress=611 " --expt-extended-lambda)
set(WARNING_SUPPRESSION_AND_OPTION_NVCC -Xcudafe "--display_error_number --diag_suppress=2915 --diag_suppress=2912 --diag_suppress=2913 --diag_suppress=111 --diag_suppress=186 --diag_suppress=611 " --expt-extended-lambda )
FILE(WRITE cuda_options "-Xcudafe \"--display_error_number --diag_suppress=2915 --diag_suppress=2914 --diag_suppress=2912 --diag_suppress=2913 --diag_suppress=111 --diag_suppress=186 --diag_suppress=611 \" --expt-extended-lambda")
else()
message(FATAL_ERROR "CUDA is incompatible, version 9.2 is only supported")
message(FATAL_ERROR "CUDA is incompatible, version 9.2 and 10.1 is only supported")
endif()
endif()
find_package(Boost 1.68.0 COMPONENTS unit_test_framework iostreams program_options)
find_package(Boost 1.68.0 COMPONENTS unit_test_framework iostreams program_options system filesystem)
find_package(MPI)
find_package(PETSc)
find_package(HDF5)
......
#!groovy
timeout(180)
{
parallel (
......@@ -123,4 +126,5 @@ parallel (
}
)
}
SUBDIRS = src images openfpm_data openfpm_io openfpm_devices openfpm_vcluster openfpm_numerics
ACLOCAL_AMFLAGS = -I m4
bin_PROGRAMS =
pdata:
cd src && make
data:
cd openfpm_data/src && make
devices:
cd openfpm_devices/src && make
vcluster:
cd openfpm_vcluster/src && make
io:
cd openfpm_io/src && make
numerics:
cd openfpm_numerics/src && make
actual_test:
cd src && make actual_test
test_pdata:
cd src && make test
test_data:
cd openfpm_data/src && make test
test_devices:
cd openfpm_devices/src && make test
test_vcluster:
cd openfpm_vcluster/src && make test
test_io:
cd openfpm_io/src && make test
test_numerics:
cd openfpm_numerics/src && make test
test: test_devices test_data test_vcluster test_pdata test_io test_numerics
.PHONY: test_pdata test_data test_devices test_vcluster test_io test_numerics
......@@ -101,7 +101,7 @@ int main(int argc, char* argv[])
openfpm_init(&argc,&argv);
// domain
Box<3,double> domain({0.0,0.0},{2.5,2.5,2.5});
Box<3,double> domain({0.0,0.0,0.0},{2.5,2.5,2.5});
// grid size
size_t sz[3] = {128,128,128};
......
......@@ -7,7 +7,7 @@ LDIR =
OBJ = main.o
%.o: %.cpp
$(CC) -O3 -g -c --std=c++11 -o $@ $< $(INCLUDE_PATH)
$(CC) -I/usr/local/cuda/include -O3 -g -c --std=c++11 -o $@ $< $(INCLUDE_PATH)
ps_cma_es: $(OBJ)
$(CC) -o $@ $^ $(CFLAGS) $(LIBS_PATH) $(LIBS)
......
......@@ -153,8 +153,8 @@ inline void mirror(vector_dist<1,double, aggregate<double,double> > & vd, vect_d
//
vd.getLastPos()[0] = 2.0 * box.getHigh(0) - vd.getPos(key)[0];
// Prolongate the initial condition
vd.template getLastProp<0>() = f_xex2(vd.getLastPos()[0]);
// no flow boundary condition
vd.template getLastProp<0>() = vd.template getProp<0>(key);
}
}
......
......@@ -10,7 +10,6 @@
*
* This example shows more in details the functionalities of **ghost_get** and **ghost_put** for a distributed vector.
*
*
* ## Inclusion ## {#e1_v_inclusion}
*
* We activate the vector_dist functionalities
......@@ -175,7 +174,7 @@ int main(int argc, char* argv[])
* Before and after. The blue arrows in the first image indicate the vector field
* for the real particles. In the second image instead the red arrow indicate the
* vector field for the real particle. The blue arrow indicate the ghosts. We can
* note that the blue arrow doea not contain the correct vector. The reason is that
* note that the blue arrow does not contain the correct vector. The reason is that
* when we used **ghost_get** we synchronized the scalar, and the tensor, but not the vector.
*
* \see \ref e1_part_ghost
......@@ -187,6 +186,34 @@ int main(int argc, char* argv[])
* <img src="http://ppmcore.mpi-cbg.de/web/images/examples/after_ghost_get.jpg"/>
* \endhtmlonly
*
* ## So ... how I have to put these ghost_get ##
*
* The first thing to do is to place the ghost in a way that the program work
* in parallel for sure. In order to do this we can do the following reasoning:
* If we have a loop over particles we distinguish two type of loops:
*
* * A loop that iterate over particles
* * A loop that iterate over particles and neighborhood particles
*
*
* If the loop is of the first type (you do not loop over the neighborhood particles)
* ghost_get is not necessary. If I am in the second case I need a ghost_get. The
* second point is which property I have to synchronize ghost_get<...>(), or more
* practically what I have to put in the ... . To answer this we have to check all
* the properties that we use from the neighborhood particles and pass it to ghost_get
* as a list. To summarize:
\code{.unparsed}
I am doing a simple loop over particles (1), or I am looping also over neighborhood particles (2)?
For the case (1) the answer is "I do not need ghost_get". For the case (2) the answer is "I need ghost_get"
if I am on the case (2) the second question is which parameters should I use ?
The answer is look at all vd.getProp<...>(b) where b is a neighborhood particle. All ... properties should appear in
ghost_get<...>()
\endcode
* This reasoning is always enough to have ghost_get function always placed correctly. For
* more fine tuning look at the options below
*
*/
......
......@@ -16,8 +16,8 @@ else
endif
ifeq ($(PROFILE),ON)
CUDA_CC=scorep --nocompiler --cuda --mpp=mpi nvcc
CUDA_CC_LINK=scorep --nocompiler --cuda --mpp=mpi nvcc
CUDA_CC=scorep --nocompiler --cuda --mpp=mpi nvcc -ccbin=mpic++
CUDA_CC_LINK=scorep --nocompiler --cuda --mpp=mpi nvcc -ccbin=mpic++
else
CUDA_CC:=$(CUDA_CC)
CUDA_CC_LINK:=$(CUDA_CC_LINK)
......
......@@ -42,7 +42,9 @@
#define PRINT_STACKTRACE
#define STOP_ON_ERROR
#define OPENMPI
#define SE_CLASS1
//#define SE_CLASS1
#define USE_LOW_REGISTER_ITERATOR
#include "Vector/vector_dist.hpp"
#include <math.h>
......@@ -97,7 +99,7 @@ const real_number MassBound = 0.0000767656;
#ifdef TEST_RUN
const real_number t_end = 0.001;
#else
const real_number t_end = 1.500;
const real_number t_end = 1.5;
#endif
// Gravity acceleration
......@@ -341,8 +343,8 @@ __global__ void calc_forces_gpu(particles_type vd, NN_type NN, real_number W_dap
// Get the position xp of the particle
Point<3,real_number> xb = vd.getPos(b);
// if (p == q) skip this particle
if (a == b) {++Np; continue;};
// if (p == q) skip this particle this condition should be done in the r^2 = 0
//if (a == b) {++Np; continue;};
unsigned int typeb = vd.getProp<type>(b);
......@@ -393,7 +395,7 @@ __global__ void calc_forces_gpu(particles_type vd, NN_type NN, real_number W_dap
template<typename CellList> inline void calc_forces(particles & vd, CellList & NN, real_number & max_visc, size_t cnt)
{
auto part = vd.getDomainIteratorGPU(64);
auto part = vd.getDomainIteratorGPU(96);
// Update the cell-list
vd.updateCellList(NN);
......@@ -856,7 +858,8 @@ int main(int argc, char* argv[])
vd.ghost_get<type,rho,Pressure,velocity>(RUN_ON_DEVICE);
auto NN = vd.getCellListGPU(2*H / 2.0);
auto NN = vd.getCellListGPU/*<CELLLIST_GPU_SPARSE<3,float>>*/(2*H / 2.0);
NN.setBoxNN(2);
timer tot_sim;
tot_sim.start();
......@@ -892,9 +895,9 @@ int main(int argc, char* argv[])
vd.map(RUN_ON_DEVICE);
// make sort
vd.make_sort(NN);
// it sort the vector (doesn not seem to produce some advantage)
// note force calculation is anyway sorted calculation
vd.make_sort(NN);
// Calculate pressure from the density
EqState(vd);
......
......@@ -701,8 +701,6 @@ template<typename VerletList> inline double calc_forces(particles & vd, VerletLi
/* if (sum1 != sum2)
{
std::cout << "PORCA TROIA: " << std::endl;
break;
}*/
......
#define CHECKFOR_POSNAN
#define CHECKFOR_PROPNAN
//#define CHECKFOR_POSNAN
//#define CHECKFOR_PROPNAN
/*!
* \page Vector_8_DEM Vector 8 Discrete element method
......@@ -530,7 +530,7 @@ int main(int argc, char* argv[])
u_ij.get(1) = parts.getProp<cpd>(p)[cidx + 1];
u_ij.get(2) = parts.getProp<cpd>(p)[cidx + 2];
Point<3,double> F_nij = sqrt(delta_ij/2/R) * (k_n*delta_ij*n_ij - gamma_t*m_eff*v_nij);
Point<3,double> F_nij = sqrt(delta_ij/2/R) * (k_n*delta_ij*n_ij - gamma_n*m_eff*v_nij);
dF_n = dF_n + F_nij;
Point<3,double> F_tij = sqrt(delta_ij/2/R) * (-k_t*u_ij - gamma_t*m_eff*v_tij);
......
......@@ -16,7 +16,7 @@
int main(int argc, char ** argv)
{
CartesianGraphFactory<2,Graph_CSR<nm_v,nm_e>> g_factory;
CartesianGraphFactory<2,Graph_CSR<nm_v<2>,nm_e>> g_factory;
// Cartesian grid
size_t sz[2] = {20,20};
......@@ -28,19 +28,19 @@ int main(int argc, char ** argv)
// Graph to decompose
Graph_CSR<nm_v,nm_e> g = g_factory.construct<nm_e::communication,NO_VERTEX_ID,float,1,0,1>(sz,box,bc);
Graph_CSR<nm_v<2>,nm_e> g = g_factory.construct<nm_e::communication,NO_VERTEX_ID,float,1,0,1>(sz,box,bc);
// Convert the graph to metis
Metis<Graph_CSR<nm_v,nm_e>> met(g,4);
Metis<Graph_CSR<nm_v<2>,nm_e>> met(g,4);
// decompose
met.decompose<nm_v::id>();
met.decompose<nm_v_id>();
// Write the decomposition
VTKWriter<Graph_CSR<nm_v,nm_e>,VTK_GRAPH> vtk(g);
VTKWriter<Graph_CSR<nm_v<2>,nm_e>,VTK_GRAPH> vtk(g);
vtk.write("Metis/vtk_partition.vtk");
}
......
......@@ -426,7 +426,7 @@ else
echo "LIBS=-lvcluster -lofpm_pdata -lofpmmemory -lparmetis -lmetis -lboost_iostreams -lboost_program_options -lhdf5 -llibhilbert $lin_alg_lib -ldl" >> example.mk
echo "LIBS_SE2=-lvcluster -lofpmmemory_se2 -lparmetis -lmetis -lboost_iostreams -lboost_program_options -lhdf5 -llibhilbert $lin_alg_lib" >> example.mk
fi
echo "INCLUDE_PATH_NVCC=-Xcompiler="-Wno-deprecated-declarations" $(cat openmp_flags) -Xcudafe \"--display_error_number --diag_suppress=2885 --diag_suppress=2886 --diag_suppress=2887 --diag_suppress=2888 --diag_suppress=186 --diag_suppress=111\" --expt-extended-lambda -I. -I$install_base/openfpm_numerics/include -I$install_base/openfpm_pdata/include/config -I$install_base/openfpm_pdata/include -I$install_base/openfpm_data/include -I$install_base/openfpm_vcluster/include -I$install_base/openfpm_io/include -I$install_base/openfpm_devices/include -I$i_dir/METIS/include -I$i_dir/PARMETIS/include -I$i_dir/BOOST/include -I$i_dir/HDF5/include -I$i_dir/LIBHILBERT/include $lin_alg_inc" >> example.mk
echo "INCLUDE_PATH_NVCC=-Xcompiler="-Wno-deprecated-declarations" $(cat openmp_flags) "$(cat cuda_options)" -I. -I$install_base/openfpm_numerics/include -I$install_base/openfpm_pdata/include/config -I$install_base/openfpm_pdata/include -I$install_base/openfpm_data/include -I$install_base/openfpm_vcluster/include -I$install_base/openfpm_io/include -I$install_base/openfpm_devices/include -I$i_dir/METIS/include -I$i_dir/PARMETIS/include -I$i_dir/BOOST/include -I$i_dir/HDF5/include -I$i_dir/LIBHILBERT/include $lin_alg_inc" >> example.mk
cp example.mk src/example.mk
cp example.mk example/example.mk
......
openfpm_data @ 3ead9a38
Subproject commit 244900099b5931924706c5e35b062c2eca72594a
Subproject commit 3ead9a385858ebc6b62da49afa1a519af51c146e
openfpm_devices @ 9f473a4f
Subproject commit 1a3dfc0a96e4ad8ab61f7f1f7cdc9127ea99041c
Subproject commit 9f473a4f9e8bd1301fa9721bdb384b15f763aec6
openfpm_io @ d7c48264
Subproject commit 10a8194fadb8009f1dc9cb22d0f118d612170267
Subproject commit d7c48264a25f8f7aded163258fb3c182a7397542
openfpm_numerics @ 45013703
Subproject commit 4fb09911107d7a63df11e3be426a940242158307
Subproject commit 450137038552dbcf28dc4379b362087da7ad8a46
openfpm_vcluster @ 1736429d
Subproject commit 15b8a504e379fb0fe44f85735ca3e9a2bd6a2f00
Subproject commit 1736429d8c02e9663f18a72fd325b116b2c4bfe6
This diff is collapsed.
/*
* grid_dist_amr_key.hpp
*
* Created on: Sep 23, 2017
* Author: i-bird
*/
#ifndef SRC_AMR_GRID_DIST_AMR_KEY_HPP_
#define SRC_AMR_GRID_DIST_AMR_KEY_HPP_
/*! \brief Amr grid distributed key
*
* \tparam dim dimensionality
*
*/
template<unsigned int dim>
class grid_dist_amr_key
{
//! actual level
size_t lvl;
//! actual position in the distributed grid
grid_dist_key_dx<dim> key;
public:
/*! \constructor
*
* \param lvl level
* \param key distributed grid key
* \param offsets to move between levels
*
*/
inline grid_dist_amr_key(size_t lvl,
grid_dist_key_dx<dim> key)
:lvl(lvl),key(key)
{}
/*! \brief Return the grid key
*
* \return the distributed key
*
*/
inline const grid_dist_key_dx<dim> & getKey() const
{
return key;
}
/*! \brief Return the grid key (as reference)
*
* \return the distributed key
*
*/
inline grid_dist_key_dx<dim> & getKeyRef()
{
return key;
}
/*! \brief Return the level
*
* \return the level
*
*/
inline size_t getLvl() const
{
return lvl;
}
/*! \brief Return the level
*
* \param lvl level to set
*
*/
inline void setLvl(size_t lvl)
{
this->lvl = lvl;
}
/*! \brief Create a new key moving the old one
*
* \param s dimension id
* \param s number of steps
*
* \return new key
*
*/
inline grid_dist_amr_key<dim> moveSpace(size_t d,size_t s)
{
return grid_dist_amr_key<dim>(lvl,key.move(d,s));
}
};
#endif /* SRC_AMR_GRID_DIST_AMR_KEY_HPP_ */
/*
* grid_amr_dist_key_iterator.hpp
*
* Created on: Sep 22, 2017
* Author: i-bird
*/
#ifndef SRC_AMR_GRID_DIST_AMR_KEY_ITERATOR_HPP_
#define SRC_AMR_GRID_DIST_AMR_KEY_ITERATOR_HPP_
#include "Vector/map_vector.hpp"
#include "Grid/Iterators/grid_dist_id_iterator.hpp"
#include "grid_dist_amr_key.hpp"
template<unsigned int dim, typename device_grid, typename device_sub_it, typename it_type = grid_dist_iterator<dim,device_grid,device_sub_it,FREE>>
class grid_dist_amr_key_iterator
{
//! Array of grid iterators
openfpm::vector<it_type> & git;
//! actual it type
struct actual_it
{
it_type & it;
};
//! Actual distributed grid iterator
it_type * a_it;
//! iterator pointer
size_t g_c;
/*! \brief from g_c increment g_c until you find a valid grid
*
*/
void selectValidGrid()
{
// When the grid has size 0 potentially all the other informations are garbage
while (g_c < git.size() && git.get(g_c).isNext() == false ) g_c++;
// get the next grid iterator
if (g_c < git.size())
{
a_it = &git.get(g_c);
}
}
public:
/*! \brief Constructor
*
* \param git vector of iterator
*
*/
grid_dist_amr_key_iterator(openfpm::vector<it_type> & git)
:git(git),g_c(0)
{
a_it = &git.get(0);
selectValidGrid();
}
//! Destructor
~grid_dist_amr_key_iterator()
{
}
/*! \brief Get the next element
*
* \return the next grid_key
*
*/
inline grid_dist_amr_key_iterator<dim,device_grid,device_sub_it,it_type> & operator++()
{
++(*a_it);
// check if a_it is at the end
if (a_it->isNext() == true)
{return *this;}
else
{
// switch to the new iterator
g_c++;
selectValidGrid();
}
return *this;
}
/*! \brief Is there a next point
*
* \return true is there is a next point
*
*/
inline bool isNext()
{
return g_c < git.size();
}
/*! \brief Return the actual AMR grid iterator point
*
*
*/
inline grid_dist_amr_key<dim> get()
{
return grid_dist_amr_key<dim>(g_c,a_it->get());
}
/*! \brief Return the actual global grid position in the AMR struct in global
* coordinates
*
*
*/
inline grid_key_dx<dim> getGKey()
{
return git.get(g_c).getGKey(a_it->get());
}
/*! \brief Return the level at which we are
*
*
*/
inline size_t getLvl() const
{
return g_c;
}
};
#endif /* SRC_AMR_GRID_DIST_AMR_KEY_ITERATOR_HPP_ */
This diff is collapsed.
/*
* amr_base_gpu_unit_tests.cu
*
* Created on: Aug 28, 2019
* Author: i-bird
*/
/*
* amr_base_unit_test.cpp
*
* Created on: Oct 5, 2017
* Author: i-bird
*/
#define BOOST_TEST_DYN_LINK
#include <boost/test/unit_test.hpp>
#include "Amr/grid_dist_amr.hpp"
#include "Point_test.hpp"
#include "Grid/tests/grid_dist_id_util_tests.hpp"
struct amr_launch_sparse
{
template<typename grid_type, typename ite_type>
__device__ void operator()(grid_type & grid, ite_type itg, float spacing, Point<3,float> center)
{
GRID_ID_3_GLOBAL(itg);
__shared__ bool is_block_empty;
if (threadIdx.x == 0 && threadIdx.y == 0 && threadIdx.z == 0)
{is_block_empty = true;}
grid.init();
int offset = 0;
grid_key_dx<3,int> blk;
bool out = grid.getInsertBlockOffset(itg,key,blk,offset);
auto blockId = grid.getBlockLinId(blk);
const float x = keyg.get(0)*spacing - center.get(0);
const float y = keyg.get(1)*spacing - center.get(1);
const float z = keyg.get(2)*spacing - center.get(2);
float radius = sqrt((float) (x*x + y*y + z*z));
bool is_active = radius < 0.4 && radius > 0.3;
if (is_active == true)
{is_block_empty = false;}
__syncthreads();
if (is_block_empty == false)
{
auto ec = grid.insertBlock(blockId);
if ( is_active == true)
{
ec.template get<0>()[offset] = x+y+z;
ec.template get<grid_type::pMask>()[offset] = 1;
}
}
__syncthreads();
grid.flush_block_insert();
}
};
BOOST_AUTO_TEST_SUITE( amr_grid_dist_id_test )
BOOST_AUTO_TEST_CASE( grid_dist_id_amr_gpu )
{
// Domain
Box<3,float> domain3({0.0,0.0,0.0},{1.0,1.0,1.0});
Ghost<3,long int> g(1);
sgrid_dist_amr_gpu<3,float,aggregate<float>> amr_g(domain3,g);
size_t g_sz[3] = {4,4,4};
size_t n_lvl = 6;
amr_g.initLevels(n_lvl,g_sz);
for (size_t i = 0 ; i < amr_g.getNLvl() ; i++)
{
// Fill the AMR with something
size_t count = 0;
auto it = amr_g.getGridIteratorGPU(i);
it.setGPUInsertBuffer(1);
Point<3,float> center({0.5,0.5,0.5});
it.launch(amr_launch_sparse(),it.getSpacing(0),center);
amr_g.getDistGrid(i).template flush<smax_<0>>(FLUSH_ON_DEVICE);
amr_g.getDistGrid(i).template deviceToHost<0>();
auto it2 = amr_g.getDistGrid(i).getDomainIterator();
while (it2.isNext())
{
auto key = it2.get();
auto keyg = it2.getGKey(key);
count++;
++it2;
}
auto & v_cl = create_vcluster();
v_cl.sum(count);
v_cl.execute();
switch(i)
{
case 0:
BOOST_REQUIRE_EQUAL(count,0);
break;
case 1:
BOOST_REQUIRE_EQUAL(count,30);
break;
case 2:
BOOST_REQUIRE_EQUAL(count,282);
break;
case 3:
BOOST_REQUIRE_EQUAL(count,2192);
break;
case 4:
BOOST_REQUIRE_EQUAL(count,16890);
break;
case 5:
BOOST_REQUIRE_EQUAL(count,136992);
break;
}
}
// Iterate across all the levels initialized
/* auto it = amr_g.getDomainIterator();
size_t count = 0;
while (it.isNext())
{
count++;
++it;
}
Vcluster<> & v_cl = create_vcluster();
v_cl.sum(count);
v_cl.execute();
BOOST_REQUIRE_EQUAL(count,correct_result);
auto itc = amr_g.getDomainIteratorCells();
size_t count_c = 0;
while (itc.isNext())
{
count_c++;
++itc;
}
v_cl.sum(count_c);
v_cl.execute();
auto it_level = amr_g.getDomainIteratorCells(3);
while (it_level.isNext())
{
auto key = it_level.get();
amr_g.template get<0>(3,key);
++it_level;
}
BOOST_REQUIRE_EQUAL(count_c,correct_result_cell);*/
}
BOOST_AUTO_TEST_SUITE_END()
/*
* amr_base_unit_test.cpp
*
* Created on: Oct 5, 2017
* Author: i-bird
*/
#define BOOST_TEST_DYN_LINK
#include <boost/test/unit_test.hpp>
#include "Grid/grid_dist_id.hpp"
#include "Point_test.hpp"
#include "Grid/tests/grid_dist_id_util_tests.hpp"
BOOST_AUTO_TEST_SUITE( amr_grid_dist_id_test )
BOOST_AUTO_TEST_CASE( grid_dist_id_amr )
{
// Domain
Box<2,float> domain2({0.0,0.0},{1.0,1.0});
size_t sz[2] = {100,100};
// Ghost
Ghost<2,long int> g(1);
// periodicity
periodicity<2> pr = {{PERIODIC,PERIODIC}};
openfpm::vector<Box<2,long int>> C_draw;
C_draw.add(Box<2,long int>({20,20},{50,24}));
C_draw.add(Box<2,long int>({51,20},{60,24}));
C_draw.add(Box<2,long int>({61,20},{70,24}));
C_draw.add(Box<2,long int>({20,25},{24,66}));
C_draw.add(Box<2,long int>({15,67},{49,85}));
C_draw.add(Box<2,long int>({50,76},{70,81}));
C_draw.add(Box<2,long int>({30,25},{34,37}));
C_draw.add(Box<2,long int>({50,66},{70,70}));
size_t volume_key = 0;
for (size_t i = 0 ; i < C_draw.size() ; i++)
{
volume_key += Box<2,long int>(C_draw.get(i)).getVolumeKey();
}
// Distributed grid with id decomposition
grid_dist_id<2,float,Point_test<float>> g_dist(sz,domain2,g,pr,C_draw);
// fill with gkey
auto git = g_dist.getDomainIterator();
grid_sm<2,void> gs(sz);
size_t count = 0;
while (git.isNext())
{
auto key = git.get();
auto gkey = git.getGKey(key);
g_dist.template get<0>(key) = gs.LinId(gkey);
count++;
++git;
}
Vcluster<> & vcl = create_vcluster();
vcl.sum(count);
vcl.execute();
BOOST_REQUIRE_EQUAL(count,volume_key);
g_dist.ghost_get<0>();
// Check it is correct
bool check = true;
size_t check_count = 0;
auto git2 = g_dist.getDomainGhostIterator();
while (git2.isNext())
{
auto key = git2.get();
auto gkey = git2.getGKey(key);
float value = g_dist.template get<0>(key);
// check if the point is inside or outside the domain
for (size_t k = 0; k < C_draw.size() ; k++)
{
if (Box<2,long int>(C_draw.get(k)).isInside(gkey.toPoint()) == true)
{
check &= value == gs.LinId(gkey);
// get the gdb_ext
auto & gdb_ext = g_dist.getLocalGridsInfo();
for (size_t s = 0 ; s < gdb_ext.size() ; s++)
{
Box<2,long int> bx = gdb_ext.get(s).Dbox;
bx += gdb_ext.get(s).origin;
if (bx.isInside(gkey.toPoint()))
{
check_count++;
break;
}
}
break;
}
}
++git2;
}
vcl.sum(check_count);
vcl.execute();
BOOST_REQUIRE_EQUAL(check,true);
BOOST_REQUIRE(check_count >= volume_key);
}
BOOST_AUTO_TEST_CASE( amr_grid_dist_id_iterator_test_use_2D)
{
// Domain
Box<2,float> domain({0.0,0.0},{1.0,1.0});
#ifdef TEST_COVERAGE_MODE
long int k = 256*256*create_vcluster().getProcessingUnits();
#else
long int k = 1024*1024*create_vcluster().getProcessingUnits();
#endif
k = std::pow(k, 1/2.);
long int big_step = k / 30;
big_step = (big_step == 0)?1:big_step;
long int small_step = 21;
print_test( "AMR Testing 2D full grid k<=",k);
// 2D test
for ( ; k >= 2 ; k-= (k > 2*big_step)?big_step:small_step )
{
BOOST_TEST_CHECKPOINT( "AMR Testing 2D full grid k=" << k );
//! [Create and access a distributed grid]
// grid size
size_t sz[2];
sz[0] = k;
sz[1] = k;
// periodicity
periodicity<2> pr = {{PERIODIC,PERIODIC}};
// Ghost
Ghost<2,long int> g(1);
openfpm::vector<Box<2,long int>> bx_def;
bx_def.add(Box<2,long int>({0,0},{k-1,k-1}));
// Distributed grid with id decomposition
grid_dist_id<2, float, aggregate<double>> g_dist(sz,domain,g,pr,bx_def);
Test2D_core(g_dist,sz,k);
}
}
BOOST_AUTO_TEST_SUITE_END()
......@@ -4,12 +4,43 @@ cmake_minimum_required(VERSION 3.8 FATAL_ERROR)
########################### Executables
if(CUDA_FOUND)
set(CUDA_SOURCES Vector/cuda/vector_dist_gpu_MP_tests.cu Vector/cuda/vector_dist_cuda_func_test.cu Decomposition/cuda/decomposition_cuda_tests.cu Vector/cuda/vector_dist_gpu_unit_tests.cu ../openfpm_devices/src/memory/CudaMemory.cu)
set(CUDA_SOURCES
Grid/tests/sgrid_dist_id_gpu_unit_tests.cu
Vector/cuda/vector_dist_gpu_MP_tests.cu
Vector/cuda/vector_dist_cuda_func_test.cu
Decomposition/cuda/decomposition_cuda_tests.cu
Vector/cuda/vector_dist_gpu_unit_tests.cu
../openfpm_devices/src/memory/CudaMemory.cu
Decomposition/cuda/Domain_icells_cart_unit_test.cu
Amr/tests/amr_base_gpu_unit_tests.cu)
else()
set(CUDA_SOURCES)
endif()
add_executable(pdata ${OPENFPM_INIT_FILE} ${CUDA_SOURCES} main.cpp Debug/debug_test.cpp Grid/tests/grid_dist_id_HDF5_chckpnt_restart_test.cpp Grid/tests/grid_dist_id_unit_test.cpp Grid/tests/staggered_grid_dist_unit_test.cpp Vector/tests/vector_dist_cell_list_tests.cpp Vector/tests/vector_dist_complex_prp_unit_test.cpp Vector/tests/vector_dist_HDF5_chckpnt_restart_test.cpp Vector/tests/vector_dist_MP_unit_tests.cpp Vector/tests/vector_dist_NN_tests.cpp Vector/tests/vector_dist_unit_test.cpp pdata_performance.cpp Decomposition/tests/CartDecomposition_unit_test.cpp Decomposition/tests/shift_vect_converter_tests.cpp Vector/performance/vector_dist_performance_util.cpp lib/pdata.cpp test_multiple_o.cpp ../openfpm_devices/src/memory/HeapMemory.cpp ../openfpm_devices/src/memory/PtrMemory.cpp ../openfpm_vcluster/src/VCluster/VCluster.cpp ../openfpm_devices/src/Memleak_check.cpp)
add_executable(pdata ${OPENFPM_INIT_FILE} ${CUDA_SOURCES} main.cpp
Amr/grid_dist_amr_unit_tests.cpp
Amr/tests/amr_base_unit_tests.cpp
Debug/debug_test.cpp
Grid/tests/grid_dist_id_HDF5_chckpnt_restart_test.cpp
Grid/tests/grid_dist_id_unit_test.cpp
Grid/tests/sgrid_dist_id_unit_tests.cpp
Grid/tests/grid_dist_id_dlb_unit_test.cpp
Grid/tests/staggered_grid_dist_unit_test.cpp
Vector/tests/vector_dist_cell_list_tests.cpp
Vector/tests/vector_dist_complex_prp_unit_test.cpp
Vector/tests/vector_dist_HDF5_chckpnt_restart_test.cpp
Vector/tests/vector_dist_MP_unit_tests.cpp
Vector/tests/vector_dist_NN_tests.cpp
Vector/tests/vector_dist_unit_test.cpp
pdata_performance.cpp
Decomposition/tests/CartDecomposition_unit_test.cpp
Decomposition/tests/shift_vect_converter_tests.cpp
Vector/performance/vector_dist_performance_util.cpp
lib/pdata.cpp test_multiple_o.cpp
../openfpm_devices/src/memory/HeapMemory.cpp
../openfpm_devices/src/memory/PtrMemory.cpp
../openfpm_vcluster/src/VCluster/VCluster.cpp
../openfpm_devices/src/Memleak_check.cpp)
if ( CMAKE_COMPILER_IS_GNUCC )
target_compile_options(pdata PRIVATE "-Wno-deprecated-declarations")
......@@ -53,12 +84,17 @@ target_include_directories (pdata PUBLIC ${LIBHILBERT_INCLUDE_DIRS})
target_include_directories (pdata PUBLIC ${Boost_INCLUDE_DIRS})
target_link_libraries(pdata ${Boost_LIBRARIES})
target_link_libraries(pdata -L${PARMETIS_ROOT}/lib parmetis)
target_link_libraries(pdata ${PARMETIS_LIBRARIES})
target_link_libraries(pdata -L${METIS_ROOT}/lib metis)
target_link_libraries(pdata -L${HDF5_ROOT}/lib hdf5 hdf5_hl)
target_link_libraries(pdata -L${LIBHILBERT_LIBRARY_DIRS} ${LIBHILBERT_LIBRARIES})
target_link_libraries(pdata ${PETSC_LIBRARIES})
if (TEST_PERFORMANCE)
target_link_libraries(pdata ${Boost_FILESYSTEM_LIBRARY})
target_link_libraries(pdata ${Boost_SYSTEM_LIBRARY})
endif()
if (TEST_COVERAGE)
target_link_libraries(pdata -lgcov --coverage)
endif()
......
This diff is collapsed.
......@@ -246,8 +246,8 @@ public:
for (size_t i = 0 ; i < dim ; i++)
this->bc[i] = dec.bc[i];
(static_cast<nn_prcs<dim,T> &>(*this)).create(this->box_nn_processor, this->sub_domains);
(static_cast<nn_prcs<dim,T> &>(*this)).applyBC(ext_domain,g,this->bc);
(static_cast<nn_prcs<dim,T,layout_base,Memory> &>(*this)).create(this->box_nn_processor, this->sub_domains);
(static_cast<nn_prcs<dim,T, layout_base,Memory> &>(*this)).applyBC(ext_domain,g,this->bc);
this->Initialize_geo_cell_lists();
this->calculateGhostBoxes();
......
......@@ -26,10 +26,10 @@ class DistParMetisDistribution
Box<dim, T> domain;
//! Processor sub-sub-domain graph
DistGraph_CSR<nm_v, nm_e> g;
DistGraph_CSR<nm_v<dim>, nm_e> g;
//! Convert the graph to parmetis format
DistParmetis<DistGraph_CSR<nm_v, nm_e>> parmetis_graph;
DistParmetis<DistGraph_CSR<nm_v<dim>, nm_e>> parmetis_graph;
//! Init vtxdist needed for Parmetis
openfpm::vector<idx_t> vtxdist;
......@@ -91,20 +91,20 @@ public:
domain = dom;
//! Create sub graph
DistGraphFactory<dim, DistGraph_CSR<nm_v, nm_e>> dist_g_factory;
DistGraphFactory<dim, DistGraph_CSR<nm_v<dim>, nm_e>> dist_g_factory;
g = dist_g_factory.template construct<NO_EDGE, T, dim - 1, 0>(gr.getSize(), domain);
g.getDecompositionVector(vtxdist);
if (dim == 2)
for (size_t i = 0; i < g.getNVertex(); i++)
g.vertex(i).template get<nm_v::x>()[2] = 0;
g.vertex(i).template get<nm_v_x>()[2] = 0;
}
/*! \brief Get the current graph (main)
*
*/
DistGraph_CSR<nm_v, nm_e> & getGraph()
DistGraph_CSR<nm_v<dim>, nm_e> & getGraph()
{
return g;
}
......@@ -118,7 +118,7 @@ public:
parmetis_graph.initSubGraph(g);
//! Decompose
parmetis_graph.decompose<nm_v::proc_id>(g);
parmetis_graph.template decompose<nm_v_proc_id>(g);
//! Get result partition for this processors
idx_t *partition = parmetis_graph.getPartition();
......@@ -143,7 +143,7 @@ public:
parmetis_graph.reset(g);
//! Refine
parmetis_graph.refine<nm_v::proc_id>(g);
parmetis_graph.template refine<nm_v_proc_id>(g);
//! Get result partition for this processors
idx_t *partition = parmetis_graph.getPartition();
......@@ -194,10 +194,10 @@ public:
std::cerr << __FILE__ << ":" << __LINE__ << " Position - Such vertex doesn't exist (id = " << id << ", " << "total size = " << g.getNVertex() << ")\n";
#endif
pos[0] = g.vertex(id).template get<nm_v::x>()[0];
pos[1] = g.vertex(id).template get<nm_v::x>()[1];
pos[0] = g.vertex(id).template get<nm_v_x>()[0];
pos[1] = g.vertex(id).template get<nm_v_x>()[1];
if (dim == 3)
pos[2] = g.vertex(id).template get<nm_v::x>()[2];
pos[2] = g.vertex(id).template get<nm_v_x>()[2];
}
/*! \brief Function that set the weight of the vertex
......@@ -215,7 +215,7 @@ public:
#endif
// If the vertex is inside this processor update the value
g.vertex(id).template get<nm_v::computation>() = weight;
g.vertex(id).template get<nm_v_computation>() = weight;
}
......@@ -242,7 +242,7 @@ public:
std::cerr << __FILE__ << ":" << __LINE__ << "Such vertex doesn't exist (id = " << id << ", " << "total size = " << g.getTotNVertex() << ")\n";
#endif
return g.vertex(id).template get<nm_v::computation>();
return g.vertex(id).template get<nm_v_computation>();
}
/*! \brief Compute the processor load counting the total weights of its vertices
......@@ -255,7 +255,7 @@ public:
for (size_t i = 0; i < g.getNVertex(); i++)
{
load += g.vertex(i).template get<nm_v::computation>();
load += g.vertex(i).template get<nm_v_computation>();
}
return load;
}
......@@ -292,7 +292,7 @@ public:
std::cerr << __FILE__ << ":" << __LINE__ << "Migration - Such vertex doesn't exist (id = " << id << ", " << "total size = " << g.getNVertex() << ")\n";
#endif
g.vertex(id).template get<nm_v::migration>() = migration;
g.vertex(id).template get<nm_v_migration>() = migration;
}
/*! \brief Set communication cost of the edge id
......@@ -338,7 +338,7 @@ public:
*/
void write(const std::string & file)
{
VTKWriter<DistGraph_CSR<nm_v, nm_e>, DIST_GRAPH> gv2(g);
VTKWriter<DistGraph_CSR<nm_v<dim>, nm_e>, DIST_GRAPH> gv2(g);
gv2.write(std::to_string(file + ".vtk"));
}
......
......@@ -53,13 +53,13 @@ class MetisDistribution
Box<dim, T> domain;
//! Global sub-sub-domain graph
Graph_CSR<nm_v, nm_e> gp;
Graph_CSR<nm_v<dim>, nm_e> gp;
//! Flag that indicate if we are doing a test (In general it fix the seed)
bool testing = false;
//! Metis decomposer utility
Metis<Graph_CSR<nm_v, nm_e>> metis_graph;
Metis<Graph_CSR<nm_v<dim>, nm_e>> metis_graph;
//! unordered map that map global sub-sub-domain to owned_cost_sub id
std::unordered_map<size_t,size_t> owner_scs;
......@@ -104,7 +104,7 @@ class MetisDistribution
public:
static constexpr unsigned int computation = nm_v::computation;
static constexpr unsigned int computation = nm_v_computation;
/*! \brief constructor
*
......@@ -182,20 +182,20 @@ public:
domain = dom;
// Create a cartesian grid graph
CartesianGraphFactory<dim, Graph_CSR<nm_v, nm_e>> g_factory_part;
gp = g_factory_part.template construct<NO_EDGE, nm_v::id, T, dim - 1, 0>(gr.getSize(), domain, bc);
CartesianGraphFactory<dim, Graph_CSR<nm_v<dim>, nm_e>> g_factory_part;
gp = g_factory_part.template construct<NO_EDGE, nm_v_id, T, dim - 1, 0>(gr.getSize(), domain, bc);
// Init to 0.0 axis z (to fix in graphFactory)
if (dim < 3)
{
for (size_t i = 0; i < gp.getNVertex(); i++)
{
gp.vertex(i).template get<nm_v::x>()[2] = 0.0;
gp.vertex(i).template get<nm_v_x>()[2] = 0.0;
}
}
for (size_t i = 0; i < gp.getNVertex(); i++)
gp.vertex(i).template get<nm_v::global_id>() = i;
gp.vertex(i).template get<nm_v_global_id>() = i;
}
/*! \brief Get the current graph (main)
......@@ -203,7 +203,7 @@ public:
* \return the current sub-sub domain Graph
*
*/
Graph_CSR<nm_v, nm_e> & getGraph()
Graph_CSR<nm_v<dim>, nm_e> & getGraph()
{
#ifdef SE_CLASS2
check_valid(this,8);
......@@ -230,7 +230,7 @@ public:
{
// we fill the assignment
for (size_t i = 0 ; i < recv_ass.size() ; i++)
gp.template vertex_p<nm_v::computation>(recv_ass.get(i).id) = recv_ass.get(i).w;
gp.template vertex_p<nm_v_computation>(recv_ass.get(i).id) = recv_ass.get(i).w;
metis_graph.initMetisGraph(v_cl.getProcessingUnits(),true);
}
......@@ -239,13 +239,13 @@ public:
metis_graph.onTest(testing);
// decompose
metis_graph.decompose<nm_v::proc_id>();
metis_graph.template decompose<nm_v_proc_id>();
if (recv_ass.size() != 0)
{
// we fill the assignment
for (size_t i = 0 ; i < recv_ass.size() ; i++)
recv_ass.get(i).w = gp.template vertex_p<nm_v::proc_id>(recv_ass.get(i).id);
recv_ass.get(i).w = gp.template vertex_p<nm_v_proc_id>(recv_ass.get(i).id);
}
else
{
......@@ -255,7 +255,7 @@ public:
for (size_t i = 0 ; i < gp.getNVertex() ; i++)
{
recv_ass.get(i).id = i;
recv_ass.get(i).w = gp.template vertex_p<nm_v::proc_id>(i);
recv_ass.get(i).w = gp.template vertex_p<nm_v_proc_id>(i);
}
}
}
......@@ -277,7 +277,7 @@ public:
// Fill the metis graph
for (size_t i = 0 ; i < recv_ass.size() ; i++)
{
gp.template vertex_p<nm_v::proc_id>(recv_ass.get(i).id) = recv_ass.get(i).w;
gp.template vertex_p<nm_v_proc_id>(recv_ass.get(i).id) = recv_ass.get(i).w;
if (recv_ass.get(i).w == v_cl.getProcessUnitID())
{
......@@ -330,10 +330,10 @@ public:
check_overflow(id);
// Copy the geometrical informations inside the pos vector
pos[0] = gp.vertex(id).template get<nm_v::x>()[0];
pos[1] = gp.vertex(id).template get<nm_v::x>()[1];
pos[0] = gp.vertex(id).template get<nm_v_x>()[0];
pos[1] = gp.vertex(id).template get<nm_v_x>()[1];
if (dim == 3)
pos[2] = gp.vertex(id).template get<nm_v::x>()[2];
{pos[2] = gp.vertex(id).template get<nm_v_x>()[2];}
}
/*! \brief function that get the computational cost of the sub-sub-domain
......@@ -349,7 +349,7 @@ public:
check_valid(this,8);
#endif
check_overflow(id);
return gp.vertex(id).template get<nm_v::computation>();
return gp.vertex(id).template get<nm_v_computation>();
}
......@@ -394,7 +394,7 @@ public:
check_overflow(id);
#endif
gp.vertex(id).template get<nm_v::migration>() = cost;
gp.vertex(id).template get<nm_v_migration>() = cost;
}
/*! \brief Set communication cost between neighborhood sub-sub-domains (weight on the edge)
......@@ -522,7 +522,7 @@ public:
check_valid(this,8);
#endif
VTKWriter<Graph_CSR<nm_v, nm_e>, VTK_GRAPH> gv2(gp);
VTKWriter<Graph_CSR<nm_v<dim>, nm_e>, VTK_GRAPH> gv2(gp);
gv2.write(std::to_string(v_cl.getProcessUnitID()) + "_" + out + ".vtk");
}
......@@ -545,7 +545,7 @@ public:
if (v_cl.getProcessUnitID() == 0)
{
for (size_t i = 0; i < gp.getNVertex(); i++)
loads.get(gp.template vertex_p<nm_v::proc_id>(i)) += gp.template vertex_p<nm_v::computation>(i);
{loads.get(gp.template vertex_p<nm_v_proc_id>(i)) += gp.template vertex_p<nm_v_computation>(i);}
for (size_t i = 0 ; i < v_cl.getProcessingUnits() ; i++)
{
......
......@@ -48,10 +48,10 @@ class ParMetisDistribution
Box<dim, T> domain;
//! Global sub-sub-domain graph
Graph_CSR<nm_v, nm_e> gp;
Graph_CSR<nm_v<dim>, nm_e> gp;
//! Convert the graph to parmetis format
Parmetis<Graph_CSR<nm_v, nm_e>> parmetis_graph;
Parmetis<Graph_CSR<nm_v<dim>, nm_e>> parmetis_graph;
//! Id of the sub-sub-domain where we set the costs
openfpm::vector<size_t> sub_sub_owner;
......@@ -115,10 +115,10 @@ class ParMetisDistribution
auto v_id = m2g.find(l)->second.id;
// Update proc id in the vertex (using the old map)
gp.template vertex_p<nm_v::proc_id>(v_id) = partitions.get(i).get(k);
gp.template vertex_p<nm_v_proc_id>(v_id) = partitions.get(i).get(k);
if (partitions.get(i).get(k) == (long int)v_cl.getProcessUnitID())
sub_sub_owner.add(v_id);
{sub_sub_owner.add(v_id);}
// Add vertex to temporary structure of distribution (needed to update main graph)
v_per_proc.get(partitions.get(i).get(k)).add(getVertexGlobalId(l));
......@@ -138,12 +138,12 @@ class ParMetisDistribution
for (size_t i = 0 ; i < gp.getNVertex(); ++i)
{
size_t pid = gp.template vertex_p<nm_v::proc_id>(i);
size_t pid = gp.template vertex_p<nm_v_proc_id>(i);
rid j = rid(vtxdist.get(pid).id + cnt.get(pid));
gid gi = gid(i);
gp.template vertex_p<nm_v::id>(i) = j.id;
gp.template vertex_p<nm_v_id>(i) = j.id;
cnt.get(pid)++;
setMapId(j,gi);
......@@ -328,8 +328,8 @@ public:
domain = dom;
// Create a cartesian grid graph
CartesianGraphFactory<dim, Graph_CSR<nm_v, nm_e>> g_factory_part;
gp = g_factory_part.template construct<NO_EDGE, nm_v::id, T, dim - 1, 0>(gr.getSize(), domain, bc);
CartesianGraphFactory<dim, Graph_CSR<nm_v<dim>, nm_e>> g_factory_part;
gp = g_factory_part.template construct<NO_EDGE