Skip to content

Commit b7d88cd

Browse files
author
dyzheng
committed
Merge branch 'tmp' of github.com:dyzheng/abacus-develop into tmp
2 parents d1caf61 + f3b1fe1 commit b7d88cd

11 files changed

Lines changed: 1176 additions & 160 deletions

File tree

‎source/module_base/module_device/cuda/memory_op.cu‎

Lines changed: 77 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -60,6 +60,17 @@ void set_memory_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_devic
6060
cudaErrcheck(cudaMemset(arr, var, sizeof(FPTYPE) * size));
6161
}
6262

63+
template <typename FPTYPE>
64+
void set_memory_2d_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_device::DEVICE_GPU* dev,
65+
FPTYPE* arr,
66+
const size_t pitch,
67+
const int var,
68+
const size_t width,
69+
const size_t height)
70+
{
71+
cudaErrcheck(cudaMemset2D(arr, sizeof(FPTYPE) * pitch , var, sizeof(FPTYPE) * width, height));
72+
}
73+
6374
template <typename FPTYPE>
6475
void synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_GPU>::operator()(
6576
const base_device::DEVICE_CPU* dev_out,
@@ -93,6 +104,48 @@ void synchronize_memory_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_
93104
cudaErrcheck(cudaMemcpy(arr_out, arr_in, sizeof(FPTYPE) * size, cudaMemcpyDeviceToDevice));
94105
}
95106

107+
template <typename FPTYPE>
108+
void synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_GPU>::operator()(
109+
const base_device::DEVICE_CPU* dev_out,
110+
const base_device::DEVICE_GPU* dev_in,
111+
FPTYPE* arr_out,
112+
const size_t dpitch,
113+
const FPTYPE* arr_in,
114+
const size_t spitch,
115+
const size_t width,
116+
const size_t height)
117+
{
118+
cudaErrcheck(cudaMemcpy2D(arr_out, dpitch * sizeof(FPTYPE), arr_in, spitch * sizeof(FPTYPE), width * sizeof(FPTYPE), height, cudaMemcpyDeviceToHost));
119+
}
120+
121+
template <typename FPTYPE>
122+
void synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_CPU>::operator()(
123+
const base_device::DEVICE_GPU* dev_out,
124+
const base_device::DEVICE_CPU* dev_in,
125+
FPTYPE* arr_out,
126+
const size_t dpitch,
127+
const FPTYPE* arr_in,
128+
const size_t spitch,
129+
const size_t width,
130+
const size_t height)
131+
{
132+
cudaErrcheck(cudaMemcpy2D(arr_out, dpitch * sizeof(FPTYPE), arr_in, spitch * sizeof(FPTYPE), width * sizeof(FPTYPE), height, cudaMemcpyHostToDevice));
133+
}
134+
135+
template <typename FPTYPE>
136+
void synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>::operator()(
137+
const base_device::DEVICE_GPU* dev_out,
138+
const base_device::DEVICE_GPU* dev_in,
139+
FPTYPE* arr_out,
140+
const size_t dpitch,
141+
const FPTYPE* arr_in,
142+
const size_t spitch,
143+
const size_t width,
144+
const size_t height)
145+
{
146+
cudaErrcheck(cudaMemcpy2D(arr_out, dpitch * sizeof(FPTYPE), arr_in, spitch * sizeof(FPTYPE), width * sizeof(FPTYPE), height, cudaMemcpyDeviceToDevice));
147+
}
148+
96149
template <typename FPTYPE_out, typename FPTYPE_in>
97150
struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_GPU, base_device::DEVICE_GPU>
98151
{
@@ -180,13 +233,20 @@ template struct resize_memory_op<float, base_device::DEVICE_GPU>;
180233
template struct resize_memory_op<double, base_device::DEVICE_GPU>;
181234
template struct resize_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
182235
template struct resize_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
236+
template struct resize_memory_op<char, base_device::DEVICE_GPU>;
183237

184238
template struct set_memory_op<int, base_device::DEVICE_GPU>;
185239
template struct set_memory_op<float, base_device::DEVICE_GPU>;
186240
template struct set_memory_op<double, base_device::DEVICE_GPU>;
187241
template struct set_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
188242
template struct set_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
189243

244+
template struct set_memory_2d_op<int, base_device::DEVICE_GPU>;
245+
template struct set_memory_2d_op<float, base_device::DEVICE_GPU>;
246+
template struct set_memory_2d_op<double, base_device::DEVICE_GPU>;
247+
template struct set_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU>;
248+
template struct set_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU>;
249+
190250
template struct synchronize_memory_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
191251
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
192252
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
@@ -203,6 +263,22 @@ template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_
203263
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
204264
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
205265

266+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
267+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
268+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
269+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
270+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
271+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
272+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
273+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
274+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
275+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
276+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
277+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
278+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
279+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
280+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
281+
206282
template struct cast_memory_op<float, float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
207283
template struct cast_memory_op<double, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
208284
template struct cast_memory_op<float, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
@@ -269,6 +345,7 @@ template struct delete_memory_op<float, base_device::DEVICE_GPU>;
269345
template struct delete_memory_op<double, base_device::DEVICE_GPU>;
270346
template struct delete_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
271347
template struct delete_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
348+
template struct delete_memory_op<char, base_device::DEVICE_GPU>;
272349
template struct delete_memory_op<float*, base_device::DEVICE_GPU>;
273350
template struct delete_memory_op<double*, base_device::DEVICE_GPU>;
274351
template struct delete_memory_op<std::complex<float>*, base_device::DEVICE_GPU>;

‎source/module_base/module_device/memory_op.cpp‎

Lines changed: 120 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -51,6 +51,18 @@ struct set_memory_op<FPTYPE, base_device::DEVICE_CPU>
5151
}
5252
};
5353

54+
template <typename FPTYPE>
55+
struct set_memory_2d_op<FPTYPE, base_device::DEVICE_CPU>
56+
{
57+
void operator()(const base_device::DEVICE_CPU* dev, FPTYPE* arr, const size_t pitch, const int var, const size_t width, const size_t height)
58+
{
59+
for (size_t i = 0; i < height; i++){
60+
set_memory_op<FPTYPE, base_device::DEVICE_CPU>()(dev, arr + i * pitch, var, width);
61+
}
62+
}
63+
};
64+
65+
5466
template <typename FPTYPE>
5567
struct synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_CPU>
5668
{
@@ -68,6 +80,25 @@ struct synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVIC
6880
}
6981
};
7082

83+
template <typename FPTYPE>
84+
struct synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_CPU>
85+
{
86+
void operator()(const base_device::DEVICE_CPU* dev_out,
87+
const base_device::DEVICE_CPU* dev_in,
88+
FPTYPE* arr_out,
89+
const size_t dpitch,
90+
const FPTYPE* arr_in,
91+
const size_t spitch,
92+
const size_t width,
93+
const size_t height)
94+
{
95+
for (int i = 0; i < height; i++){
96+
synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_CPU>()(dev_out, dev_in,
97+
arr_out + i * dpitch, arr_in + i * spitch, width);
98+
}
99+
}
100+
};
101+
71102
template <typename FPTYPE_out, typename FPTYPE_in>
72103
struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_CPU, base_device::DEVICE_CPU>
73104
{
@@ -108,12 +139,24 @@ template struct set_memory_op<double, base_device::DEVICE_CPU>;
108139
template struct set_memory_op<std::complex<float>, base_device::DEVICE_CPU>;
109140
template struct set_memory_op<std::complex<double>, base_device::DEVICE_CPU>;
110141

142+
template struct set_memory_2d_op<int, base_device::DEVICE_CPU>;
143+
template struct set_memory_2d_op<float, base_device::DEVICE_CPU>;
144+
template struct set_memory_2d_op<double, base_device::DEVICE_CPU>;
145+
template struct set_memory_2d_op<std::complex<float>, base_device::DEVICE_CPU>;
146+
template struct set_memory_2d_op<std::complex<double>, base_device::DEVICE_CPU>;
147+
111148
template struct synchronize_memory_op<int, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
112149
template struct synchronize_memory_op<float, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
113150
template struct synchronize_memory_op<double, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
114151
template struct synchronize_memory_op<std::complex<float>, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
115152
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
116153

154+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
155+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
156+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
157+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
158+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
159+
117160
template struct cast_memory_op<float, float, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
118161
template struct cast_memory_op<double, double, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
119162
template struct cast_memory_op<float, double, base_device::DEVICE_CPU, base_device::DEVICE_CPU>;
@@ -165,6 +208,13 @@ struct set_memory_op<FPTYPE, base_device::DEVICE_GPU>
165208
{
166209
}
167210
};
211+
template <typename FPTYPE>
212+
struct set_memory_2d_op<FPTYPE, base_device::DEVICE_GPU>
213+
{
214+
void operator()(const base_device::DEVICE_GPU* dev, FPTYPE* arr, const size_t pitch, const int var, const size_t width, const size_t height)
215+
{
216+
}
217+
};
168218

169219
template <typename FPTYPE>
170220
struct synchronize_memory_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>
@@ -202,6 +252,54 @@ struct synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVIC
202252
}
203253
};
204254

255+
template <typename FPTYPE>
256+
struct synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>
257+
{
258+
void operator()(const base_device::DEVICE_GPU* dev_out,
259+
const base_device::DEVICE_GPU* dev_in,
260+
const Device_in* dev_in,
261+
FPTYPE* arr_out,
262+
const size_t dpitch,
263+
const FPTYPE* arr_in,
264+
const size_t spitch,
265+
const size_t width,
266+
const size_t height)
267+
{
268+
}
269+
};
270+
271+
template <typename FPTYPE>
272+
struct synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_CPU>
273+
{
274+
void operator()(const base_device::DEVICE_GPU* dev_out,
275+
const base_device::DEVICE_CPU* dev_in,
276+
const Device_in* dev_in,
277+
FPTYPE* arr_out,
278+
const size_t dpitch,
279+
const FPTYPE* arr_in,
280+
const size_t spitch,
281+
const size_t width,
282+
const size_t height)
283+
{
284+
}
285+
};
286+
287+
template <typename FPTYPE>
288+
struct synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_GPU>
289+
{
290+
void operator()(const base_device::DEVICE_CPU* dev_out,
291+
const base_device::DEVICE_GPU* dev_in,
292+
const Device_in* dev_in,
293+
FPTYPE* arr_out,
294+
const size_t dpitch,
295+
const FPTYPE* arr_in,
296+
const size_t spitch,
297+
const size_t width,
298+
const size_t height)
299+
{
300+
}
301+
};
302+
205303
template <typename FPTYPE_out, typename FPTYPE_in>
206304
struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_GPU, base_device::DEVICE_GPU>
207305
{
@@ -258,6 +356,12 @@ template struct set_memory_op<double, base_device::DEVICE_GPU>;
258356
template struct set_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
259357
template struct set_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
260358

359+
template struct set_memory_2d_op<int, base_device::DEVICE_GPU>;
360+
template struct set_memory_2d_op<float, base_device::DEVICE_GPU>;
361+
template struct set_memory_2d_op<double, base_device::DEVICE_GPU>;
362+
template struct set_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU>;
363+
template struct set_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU>;
364+
261365
template struct synchronize_memory_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
262366
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
263367
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
@@ -274,6 +378,22 @@ template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_
274378
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
275379
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
276380

381+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
382+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
383+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
384+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
385+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
386+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
387+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
388+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
389+
template struct synchronize_memory_2d_op<double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
390+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
391+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
392+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
393+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
394+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
395+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
396+
277397
template struct cast_memory_op<float, float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
278398
template struct cast_memory_op<double, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
279399
template struct cast_memory_op<float, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;

0 commit comments

Comments
 (0)