Shamrock 2025.10.0
Astrophysical Code
Loading...
Searching...
No Matches
TreeStructure.cpp
Go to the documentation of this file.
1// -------------------------------------------------------//
2//
3// SHAMROCK code for hydrodynamics
4// Copyright (c) 2021-2026 Timothée David--Cléris <tim.shamrock@proton.me>
5// SPDX-License-Identifier: CeCILL Free Software License Agreement v2.1
6// Shamrock is licensed under the CeCILL 2.1 License, see LICENSE for more information
7//
8// -------------------------------------------------------//
9
15
16#include "shamtree/TreeStructure.hpp"
17#include "shamalgs/details/memory/memory.hpp"
18#include "shamalgs/primitives/equals.hpp"
19#include "shamsys/NodeInstance.hpp"
20#include "shamtree/kernels/karras_alg.hpp"
21
22namespace shamrock::tree {
23
24 template<class u_morton>
25 void TreeStructure<u_morton>::build(
26 sycl::queue &queue, u32 _internal_cell_count, sycl::buffer<u_morton> &morton_buf) {
27
28 if (!(_internal_cell_count < morton_buf.size())) {
29 throw shambase::make_except_with_loc<std::runtime_error>(
30 "morton buf must be at least with size() greater than internal_cell_count");
31 }
32
33 internal_cell_count = _internal_cell_count;
34
35 buf_lchild_id = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
36 buf_rchild_id = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
37 buf_lchild_flag = std::make_unique<sycl::buffer<u8>>(internal_cell_count);
38 buf_rchild_flag = std::make_unique<sycl::buffer<u8>>(internal_cell_count);
39 buf_endrange = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
40
41 sycl_karras_alg(
42 queue,
43 internal_cell_count,
44 morton_buf,
45 *buf_lchild_id,
46 *buf_rchild_id,
47 *buf_lchild_flag,
48 *buf_rchild_flag,
49 *buf_endrange);
50
51 one_cell_mode = false;
52 }
53
54 template<class u_morton>
55 void TreeStructure<u_morton>::build_one_cell_mode() {
56 internal_cell_count = 1;
57 buf_lchild_id = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
58 buf_rchild_id = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
59 buf_lchild_flag = std::make_unique<sycl::buffer<u8>>(internal_cell_count);
60 buf_rchild_flag = std::make_unique<sycl::buffer<u8>>(internal_cell_count);
61 buf_endrange = std::make_unique<sycl::buffer<u32>>(internal_cell_count);
62
63 {
64 sycl::host_accessor rchild_id{*buf_rchild_id, sycl::write_only, sycl::no_init};
65 sycl::host_accessor lchild_id{*buf_lchild_id, sycl::write_only, sycl::no_init};
66 sycl::host_accessor rchild_flag{*buf_rchild_flag, sycl::write_only, sycl::no_init};
67 sycl::host_accessor lchild_flag{*buf_lchild_flag, sycl::write_only, sycl::no_init};
68 sycl::host_accessor endrange{*buf_endrange, sycl::write_only, sycl::no_init};
69
70 lchild_id[0] = 0;
71 rchild_id[0] = 1;
72 lchild_flag[0] = 1;
73 rchild_flag[0] = 1;
74
75 endrange[0] = 1;
76 }
77 one_cell_mode = true;
78 }
79
80 template<class u_morton>
81 TreeStructure<u_morton>::TreeStructure(const TreeStructure<u_morton> &other)
82 : internal_cell_count(other.internal_cell_count), one_cell_mode(other.one_cell_mode),
83 buf_lchild_id(
84 shamalgs::memory::duplicate(
85 shamsys::instance::get_compute_queue(),
86 other.buf_lchild_id)), // size = internal
87 buf_rchild_id(
88 shamalgs::memory::duplicate(
89 shamsys::instance::get_compute_queue(),
90 other.buf_rchild_id)), // size = internal
91 buf_lchild_flag(
92 shamalgs::memory::duplicate(
93 shamsys::instance::get_compute_queue(),
94 other.buf_lchild_flag)), // size = internal
95 buf_rchild_flag(
96 shamalgs::memory::duplicate(
97 shamsys::instance::get_compute_queue(),
98 other.buf_rchild_flag)), // size = internal
99 buf_endrange(
100 shamalgs::memory::duplicate(
101 shamsys::instance::get_compute_queue(),
102 other.buf_endrange)) // size = internal
103 {}
104
105 template<class u_morton>
106 bool TreeStructure<u_morton>::operator==(const TreeStructure<u_morton> &rhs) const {
107 bool cmp = true;
108
109 cmp = cmp && (internal_cell_count == rhs.internal_cell_count);
110
111 cmp = cmp
112 && shamalgs::primitives::equals(
113 shamsys::instance::get_compute_queue(),
114 *buf_lchild_id,
115 *rhs.buf_lchild_id,
116 internal_cell_count);
117 cmp = cmp
118 && shamalgs::primitives::equals(
119 shamsys::instance::get_compute_queue(),
120 *buf_rchild_id,
121 *rhs.buf_rchild_id,
122 internal_cell_count);
123 cmp = cmp
124 && shamalgs::primitives::equals(
125 shamsys::instance::get_compute_queue(),
126 *buf_lchild_flag,
127 *rhs.buf_lchild_flag,
128 internal_cell_count);
129 cmp = cmp
130 && shamalgs::primitives::equals(
131 shamsys::instance::get_compute_queue(),
132 *buf_rchild_flag,
133 *rhs.buf_rchild_flag,
134 internal_cell_count);
135 cmp = cmp
136 && shamalgs::primitives::equals(
137 shamsys::instance::get_compute_queue(),
138 *buf_endrange,
139 *rhs.buf_endrange,
140 internal_cell_count);
141 cmp = cmp && (one_cell_mode == rhs.one_cell_mode);
142
143 return cmp;
144 }
145
146 template class TreeStructure<u32>;
147 template class TreeStructure<u64>;
148
149} // namespace shamrock::tree
NodeInstance.hpp
Header file describing a Node Instance.
shamsys::instance::get_compute_queue
sycl::queue & get_compute_queue(u32 id=0)
Definition NodeInstance.cpp:370
u32
std::uint32_t u32
32 bit unsigned integer
Definition aliases_int.hpp:27
equals.hpp
Element-wise equality comparison algorithms for buffers.
karras_alg.hpp
sycl_karras_alg
void sycl_karras_alg(sycl::queue &queue, u32 internal_cell_count, sycl::buffer< u_morton > &in_morton, sycl::buffer< u32 > &out_buf_lchild_id, sycl::buffer< u32 > &out_buf_rchild_id, sycl::buffer< u8 > &out_buf_lchild_flag, sycl::buffer< u8 > &out_buf_rchild_flag, sycl::buffer< u32 > &out_buf_endrange)
Karras 2012 algorithm with addition endrange buffer.
shamalgs::primitives::equals
bool equals(sycl::queue &q, sycl::buffer< T > &buf1, sycl::buffer< T > &buf2, u32 cnt)
Compare elements between two sycl::buffers for equality.
Definition equals.hpp:77
shambase::make_except_with_loc
ExcptTypes make_except_with_loc(std::string message, SourceLocation loc=SourceLocation{})
Create an exception with a message and a location.
Definition exception.hpp:105
Generated on for Shamrock by doxygen 1.15.0