1 | /* |
2 | * CellDecomposer_gpu_ker.hpp |
3 | * |
4 | * Created on: Apr 28, 2019 |
5 | * Author: i-bird |
6 | */ |
7 | |
8 | #ifndef CELLDECOMPOSER_GPU_KER_HPP_ |
9 | #define CELLDECOMPOSER_GPU_KER_HPP_ |
10 | |
11 | #include "util/cuda_launch.hpp" |
12 | #include "util/multi_array_openfpm/array_openfpm.hpp" |
13 | #include "Grid/grid_sm.hpp" |
14 | #include "NN/CellList/cuda/Cuda_cell_list_util_func.hpp" |
15 | #include "NN/CellList/CellDecomposer.hpp" |
16 | |
17 | template <unsigned int dim, typename T, typename cnt_type, typename ids_type, typename transform> |
18 | class CellDecomposer_gpu_ker |
19 | { |
20 | //! Spacing |
21 | openfpm::array<T,dim,cnt_type> spacing_c; |
22 | |
23 | //! \brief number of sub-divisions in each direction |
24 | openfpm::array<ids_type,dim,cnt_type> div_c; |
25 | |
26 | //! \brief cell offset |
27 | openfpm::array<ids_type,dim,cnt_type> off; |
28 | |
29 | //! transformation |
30 | transform t; |
31 | |
32 | public: |
33 | |
34 | __device__ __host__ CellDecomposer_gpu_ker() |
35 | {} |
36 | |
37 | __device__ __host__ CellDecomposer_gpu_ker(openfpm::array<T,dim,cnt_type> & spacing_c, |
38 | openfpm::array<ids_type,dim,cnt_type> & div_c, |
39 | openfpm::array<ids_type,dim,cnt_type> & off, |
40 | const transform & t) |
41 | :spacing_c(spacing_c),div_c(div_c),off(off),t(t) |
42 | {} |
43 | |
44 | __host__ grid_sm<dim,void> getGrid() |
45 | { |
46 | size_t sz[dim]; |
47 | |
48 | for (size_t i = 0 ; i < dim ; i++) |
49 | { |
50 | sz[i] = div_c[i] + 2*off[i]; |
51 | } |
52 | |
53 | return grid_sm<dim,void> (sz); |
54 | } |
55 | |
56 | __device__ __host__ inline grid_key_dx<dim,ids_type> getCell(const Point<dim,T> & xp) const |
57 | { |
58 | return cid_<dim,cnt_type,ids_type,transform>::get_cid_key(spacing_c,off,t,xp); |
59 | } |
60 | |
61 | __device__ __host__ inline cnt_type LinId(const grid_key_dx<dim,ids_type> & k) const |
62 | { |
63 | return cid_<dim,cnt_type,ids_type,transform>::get_cid(div_c,k); |
64 | } |
65 | |
66 | __device__ inline const openfpm::array<T,dim,cnt_type> & get_spacing_c() const |
67 | { |
68 | return spacing_c; |
69 | } |
70 | |
71 | __device__ __host__ inline const openfpm::array<ids_type,dim,cnt_type> & get_div_c() const |
72 | { |
73 | return div_c; |
74 | } |
75 | |
76 | __device__ __host__ inline const openfpm::array<ids_type,dim,cnt_type> & get_off() const |
77 | { |
78 | return off; |
79 | } |
80 | |
81 | __device__ __host__ inline const transform & get_t() const |
82 | { |
83 | return t; |
84 | } |
85 | }; |
86 | |
87 | #endif /* CELLDECOMPOSER_GPU_KER_HPP_ */ |
88 | |