Skip to content

Commit 7424205

Browse files
kunkunblueberrykunkunblueberry
committed
fix the undefined function of template~
1 parent fa37449 commit 7424205

6 files changed

Lines changed: 91 additions & 154 deletions

File tree

‎source/module_base/module_device/CMakeLists.txt‎

Lines changed: 39 additions & 39 deletions
Original file line numberDiff line numberDiff line change
@@ -1,44 +1,44 @@
1-
# list(APPEND module_device_srcs
2-
# memory_op.cpp
3-
# device.cpp
4-
# )
1+
list(APPEND module_device_srcs
2+
memory_op.cpp
3+
device.cpp
4+
)
55

6-
# if(USE_CUDA)
7-
# list(APPEND module_device_srcs
8-
# cuda/memory_op.cu
9-
# )
10-
# endif()
6+
if(USE_CUDA)
7+
list(APPEND module_device_srcs
8+
cuda/memory_op.cu
9+
)
10+
endif()
1111

1212

13-
# if(USE_ROCM)
14-
# hip_add_library(module_device_rocm STATIC
15-
# rocm/memory_op.hip.cu
16-
# )
17-
# target_link_libraries(
18-
# device
19-
# module_device_rocm
20-
# hip::host
21-
# hip::device
22-
# hip::hipfft
23-
# roc::hipblas
24-
# roc::hipsolver
25-
# )
26-
# endif()
13+
if(USE_ROCM)
14+
hip_add_library(module_device_rocm STATIC
15+
rocm/memory_op.hip.cu
16+
)
17+
target_link_libraries(
18+
device
19+
module_device_rocm
20+
hip::host
21+
hip::device
22+
hip::hipfft
23+
roc::hipblas
24+
roc::hipsolver
25+
)
26+
endif()
2727

28-
# add_library(device OBJECT ${device_srcs})
28+
add_library(device OBJECT ${module_device_srcs})
2929

30-
# if(USE_CUDA)
31-
# target_link_libraries(
32-
# device
33-
# )
34-
# elseif(USE_ROCM)
35-
# target_link_libraries(
36-
# device
37-
# device_rocm
38-
# hip::host
39-
# hip::device
40-
# hip::hipfft
41-
# roc::hipblas
42-
# roc::hipsolver
43-
# )
44-
# endif()
30+
if(USE_CUDA)
31+
target_link_libraries(
32+
device
33+
)
34+
elseif(USE_ROCM)
35+
target_link_libraries(
36+
device
37+
device_rocm
38+
hip::host
39+
hip::device
40+
hip::hipfft
41+
roc::hipblas
42+
roc::hipsolver
43+
)
44+
endif()

‎source/module_base/module_device/memory_op.cpp‎

Lines changed: 5 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -346,6 +346,7 @@ template struct resize_memory_op<float, base_device::DEVICE_GPU>;
346346
template struct resize_memory_op<double, base_device::DEVICE_GPU>;
347347
template struct resize_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
348348
template struct resize_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
349+
template struct resize_memory_op<char, base_device::DEVICE_GPU>;
349350

350351
template struct set_memory_op<int, base_device::DEVICE_GPU>;
351352
template struct set_memory_op<float, base_device::DEVICE_GPU>;
@@ -457,7 +458,10 @@ template struct delete_memory_op<float, base_device::DEVICE_GPU>;
457458
template struct delete_memory_op<double, base_device::DEVICE_GPU>;
458459
template struct delete_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
459460
template struct delete_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
461+
template struct delete_memory_op<char, base_device::DEVICE_GPU>;
460462
#endif
461463

462464
} // namespace memory
463-
} // namespace base_device
465+
} // namespace base_device
466+
467+

‎source/module_base/module_device/memory_op.h‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -356,4 +356,4 @@ using castmem_z2c_d2h_op = base_device::memory::
356356
static base_device::DEVICE_CPU* cpu_ctx = {};
357357
static base_device::DEVICE_GPU* gpu_ctx = {};
358358

359-
#endif // MODULE_DEVICE_MEMORY_H_
359+
#endif // MODULE_DEVICE_MEMORY_H_

‎source/module_base/module_device/rocm/.memory_op.hip.cu.swp‎

24 KB
Binary file not shown.

‎source/module_base/module_device/rocm/memory_op.hip.cu‎

Lines changed: 41 additions & 108 deletions
Original file line numberDiff line numberDiff line change
@@ -57,9 +57,13 @@ void set_memory_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_devic
5757
const int var,
5858
const size_t size)
5959
{
60-
hipErrcheck(hipMemset(arr, var, sizeof(FPTYPE) * size));
60+
hipErrcheck(hipMemset(arr, const size_t height)
61+
{
62+
hipErrcheck(hipMemset2D(arr, sizeof(FPTYPE) * pitch , var, sizeof(FPTYPE) * width, height));
6163
}
6264

65+
66+
6367
template <typename FPTYPE>
6468
void synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_GPU>::operator()(
6569
const base_device::DEVICE_CPU* dev_out,
@@ -68,33 +72,34 @@ void synchronize_memory_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_
6872
const FPTYPE* arr_in,
6973
const size_t size)
7074
{
71-
hipErrcheck(hipMemcpy(arr_out, arr_in, sizeof(FPTYPE) * size, hipMemcpyDeviceToHost));
72-
}
73-
74-
template <typename FPTYPE>
75-
void synchronize_memory_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_CPU>::operator()(
75+
hipErrcheck(hipMemcpy(arr_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>::operator()(
7676
const base_device::DEVICE_GPU* dev_out,
77-
const base_device::DEVICE_CPU* dev_in,
77+
const base_device::DEVICE_GPU* dev_in,
7878
FPTYPE* arr_out,
7979
const FPTYPE* arr_in,
8080
const size_t size)
8181
{
82-
hipErrcheck(hipMemcpy(arr_out, arr_in, sizeof(FPTYPE) * size, hipMemcpyHostToDevice));
82+
hipErrcheck(hipMemcpy(arr_out, arr_in, sizeof(FPTYPE) * size, hipMemcpyDeviceToDevice));
8383
}
8484

8585
template <typename FPTYPE>
86-
void synchronize_memory_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>::operator()(
87-
const base_device::DEVICE_GPU* dev_out,
88-
const base_device::DEVICE_GPU* dev_in,
86+
void synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_CPU, base_device::DEVICE_GPU>::operator()(
87+
const base_device::DEVICE_CPU* dev_out,
88+
const basICE_GPU* dev_out,
89+
const base_device::DEVICE_CPU* dev_in,
8990
FPTYPE* arr_out,
91+
const size_t dpitch,
9092
const FPTYPE* arr_in,
91-
const size_t size)
93+
const size_t spitch,
94+
const size_t width,
95+
const size_t height)
9296
{
93-
hipErrcheck(hipMemcpy(arr_out, arr_in, sizeof(FPTYPE) * size, hipMemcpyDeviceToDevice));
97+
hipErrcheck(hipMemcpy2D(arr_out, dpitch * sizeof(FPTYPE), arr_in, spitch * sizeof(FPTYPE), width * sizeof(FPTYPE), height, hipMemcpyHostToDevice));
9498
}
9599

96-
template <typename FPTYPE_out, typename FPTYPE_in>
97-
struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_GPU, base_device::DEVICE_GPU> {
100+
template <typename FPTYPE>
101+
void synchronize_memory_2d_op<FPTYPE, base_device::DEVICE_GPU, base_device::DEVICE_GPU>::operator()(
102+
const base_device::DEVI_GPU, base_device::DEVICE_GPU> {
98103
void operator()(const base_device::DEVICE_GPU* dev_out,
99104
const base_device::DEVICE_GPU* dev_in,
100105
FPTYPE_out* arr_out,
@@ -104,34 +109,13 @@ struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_GPU, base_devic
104109
if (size == 0) {return;}
105110
const int block = (size + THREADS_PER_BLOCK - 1) / THREADS_PER_BLOCK;
106111
hipLaunchKernelGGL(cast_memory, dim3(block), dim3(THREADS_PER_BLOCK), 0, 0, arr_out, arr_in, size);
107-
hipCheckOnDebug();
108-
}
109-
};
110-
111-
template <typename FPTYPE_out, typename FPTYPE_in>
112-
struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_GPU, base_device::DEVICE_CPU> {
113-
void operator()(const base_device::DEVICE_GPU* dev_out,
114-
const base_device::DEVICE_CPU* dev_in,
115-
FPTYPE_out* arr_out,
116-
const FPTYPE_in* arr_in,
117-
const size_t size) {
118-
119-
if (size == 0) {return;}
120-
// No need to cast the memory if the data types are the same.
112+
hipCheckOnDebug();y if the data types are the same.
121113
if (std::is_same<FPTYPE_out, FPTYPE_in>::value)
122114
{
123115
synchronize_memory_op<FPTYPE_out, base_device::DEVICE_GPU, base_device::DEVICE_CPU>()(dev_out,
124116
dev_in,
125117
arr_out,
126-
reinterpret_cast<const FPTYPE_out*>(arr_in),
127-
size);
128-
return;
129-
}
130-
FPTYPE_in * arr = nullptr;
131-
hipErrcheck(hipMalloc((void **)&arr, sizeof(FPTYPE_in) * size));
132-
hipErrcheck(hipMemcpy(arr, arr_in, sizeof(FPTYPE_in) * size, hipMemcpyHostToDevice));
133-
const int block = (size + THREADS_PER_BLOCK - 1) / THREADS_PER_BLOCK;
134-
hipLaunchKernelGGL(cast_memory, dim3(block), dim3(THREADS_PER_BLOCK), 0, 0, arr_out, arr, size);
118+
reinterpret_clGGL(cast_memory, dim3(block), dim3(THREADS_PER_BLOCK), 0, 0, arr_out, arr, size);
135119
hipCheckOnDebug();
136120
hipErrcheck(hipFree(arr));
137121
}
@@ -143,47 +127,23 @@ struct cast_memory_op<FPTYPE_out, FPTYPE_in, base_device::DEVICE_CPU, base_devic
143127
const base_device::DEVICE_GPU* dev_in,
144128
FPTYPE_out* arr_out,
145129
const FPTYPE_in* arr_in,
146-
const size_t size) {
147-
148-
if (size == 0) {return;}
149-
// No need to cast the memory if the data types are the same.
150-
if (std::is_same<FPTYPE_out, FPTYPE_in>::value)
151-
{
152-
synchronize_memory_op<FPTYPE_out, base_device::DEVICE_CPU, base_device::DEVICE_GPU>()(dev_out,
153-
dev_in,
154-
arr_out,
130+
const siz arr_out,
155131
reinterpret_cast<const FPTYPE_out*>(arr_in),
156132
size);
157133
return;
158134
}
159135
auto * arr = (FPTYPE_in*) malloc(sizeof(FPTYPE_in) * size);
160136
hipErrcheck(hipMemcpy(arr, arr_in, sizeof(FPTYPE_in) * size, hipMemcpyDeviceToHost));
161137
for (int ii = 0; ii < size; ii++) {
162-
arr_out[ii] = static_cast<FPTYPE_out>(arr[ii]);
163-
}
164-
free(arr);
165-
}
166-
};
167-
168-
template <typename FPTYPE>
169-
void delete_memory_op<FPTYPE, base_device::DEVICE_GPU>::operator()(const base_device::DEVICE_GPU* dev, FPTYPE* arr)
170-
{
171-
hipErrcheck(hipFree(arr));
172-
}
173-
174-
template struct resize_memory_op<int, base_device::DEVICE_GPU>;
175-
template struct resize_memory_op<float, base_device::DEVICE_GPU>;
176-
template struct resize_memory_op<double, base_device::DEVICE_GPU>;
177-
template struct resize_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
138+
arr_out[mplate struct resize_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
178139
template struct resize_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
140+
template struct resize_memory_op<char, base_device::DEVICE_GPU>;
179141

180142
template struct set_memory_op<int, base_device::DEVICE_GPU>;
181143
template struct set_memory_op<float, base_device::DEVICE_GPU>;
182144
template struct set_memory_op<double, base_device::DEVICE_GPU>;
183145
template struct set_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
184-
template struct set_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
185-
186-
template struct synchronize_memory_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
146+
template struct set_memory_op<std::complex<double>, base_dee_device::DEVICE_GPU>;
187147
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
188148
template struct synchronize_memory_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
189149
template struct synchronize_memory_op<float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
@@ -199,12 +159,16 @@ template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_
199159
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
200160
template struct synchronize_memory_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
201161

202-
template struct cast_memory_op<float, float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
203-
template struct cast_memory_op<double, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
204-
template struct cast_memory_op<float, double, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
205-
template struct cast_memory_op<double, float, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
206-
template struct cast_memory_op<std::complex<float>,
207-
std::complex<float>,
162+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
163+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
164+
template struct synchronize_memory_2d_op<int, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
165+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
166+
template struct synchronize_memory_2d_op<float, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
167+
template struct synchronize_memory_2d_op<float, base_devic:complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
168+
template struct synchronize_memory_2d_op<std::complex<float>, base_device::DEVICE_GPU, base_device::DEVICE_GPU>;
169+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
170+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
171+
template struct synchronize_memory_2d_op<std::complex<double>, base_device::DEVICE_GPU, base_devic:complex<float>,
208172
base_device::DEVICE_GPU,
209173
base_device::DEVICE_GPU>;
210174
template struct cast_memory_op<std::complex<double>,
@@ -214,46 +178,21 @@ template struct cast_memory_op<std::complex<double>,
214178
template struct cast_memory_op<std::complex<float>,
215179
std::complex<double>,
216180
base_device::DEVICE_GPU,
217-
base_device::DEVICE_GPU>;
218-
template struct cast_memory_op<std::complex<double>,
219-
std::complex<float>,
220-
base_device::DEVICE_GPU,
221-
base_device::DEVICE_GPU>;
222-
template struct cast_memory_op<float, float, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
223-
template struct cast_memory_op<double, double, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
224-
template struct cast_memory_op<float, double, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
181+
memory_op<float, double, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
225182
template struct cast_memory_op<double, float, base_device::DEVICE_GPU, base_device::DEVICE_CPU>;
226183
template struct cast_memory_op<std::complex<float>,
227184
std::complex<float>,
228185
base_device::DEVICE_GPU,
229186
base_device::DEVICE_CPU>;
230187
template struct cast_memory_op<std::complex<double>,
231188
std::complex<double>,
232-
base_device::DEVICE_GPU,
233-
base_device::DEVICE_CPU>;
234-
template struct cast_memory_op<std::complex<float>,
235-
std::complex<double>,
236-
base_device::DEVICE_GPU,
237-
base_device::DEVICE_CPU>;
238-
template struct cast_memory_op<std::complex<double>,
239-
std::complex<float>,
240-
base_device::DEVICE_GPU,
241-
base_device::DEVICE_CPU>;
189+
base_device::DEVICE_CPU>;
242190
template struct cast_memory_op<float, float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
243191
template struct cast_memory_op<double, double, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
244192
template struct cast_memory_op<float, double, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
245193
template struct cast_memory_op<double, float, base_device::DEVICE_CPU, base_device::DEVICE_GPU>;
246194
template struct cast_memory_op<std::complex<float>,
247-
std::complex<float>,
248-
base_device::DEVICE_CPU,
249-
base_device::DEVICE_GPU>;
250-
template struct cast_memory_op<std::complex<double>,
251-
std::complex<double>,
252-
base_device::DEVICE_CPU,
253-
base_device::DEVICE_GPU>;
254-
template struct cast_memory_op<std::complex<float>,
255-
std::complex<double>,
256-
base_device::DEVICE_CPU,
195+
std::comp base_device::DEVICE_CPU,
257196
base_device::DEVICE_GPU>;
258197
template struct cast_memory_op<std::complex<double>,
259198
std::complex<float>,
@@ -263,12 +202,6 @@ template struct cast_memory_op<std::complex<double>,
263202
template struct delete_memory_op<int, base_device::DEVICE_GPU>;
264203
template struct delete_memory_op<float, base_device::DEVICE_GPU>;
265204
template struct delete_memory_op<double, base_device::DEVICE_GPU>;
266-
template struct delete_memory_op<std::complex<float>, base_device::DEVICE_GPU>;
267-
template struct delete_memory_op<std::complex<double>, base_device::DEVICE_GPU>;
268-
template struct delete_memory_op<float*, base_device::DEVICE_GPU>;
269-
template struct delete_memory_op<double*, base_device::DEVICE_GPU>;
270-
template struct delete_memory_op<std::complex<float>*, base_device::DEVICE_GPU>;
271-
template struct delete_memory_op<std::complex<double>*, base_device::DEVICE_GPU>;
272-
205+
template struct delete_memor base_device::DEVICE_GPU>;
273206
} // namespace memory
274-
} // end of namespace base_device
207+
} // end of namespace base_device

‎source/module_basis/module_pw/fft.cpp‎

Lines changed: 5 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -1,10 +1,11 @@
11
#include "fft.h"
2-
2+
#include "module_base/module_device/memory_op.h"
33
#include "module_base/memory.h"
44
#include "module_base/tool_quit.h"
55
#include "module_hamilt_pw/hamilt_pwdft/global.h"
66
#include "module_parameter/parameter.h"
77

8+
89
namespace ModulePW
910
{
1011

@@ -987,10 +988,10 @@ void BatchedFFT<FPTYPE>::fft3D_backward(const base_device::DEVICE_GPU* /*ctx*/,
987988
hipfftHandle plan = this->get_plan_from_cache(batchSize);
988989
if (this->fftType == HIPFFT_C2C){
989990
CHECK_CUFFT(hipfftExecC2C(plan, reinterpret_cast<hipfftComplex*>(in), reinterpret_cast<hipfftComplex*>(out),
990-
HIPFFT_INVERSE));
991+
HIPFFT_BACKWARD));
991992
}else{
992-
CHECK_CUFFT(cufftExecZ2Z(plan, reinterpret_cast<hipfftDoubleComplex*>(in), reinterpret_cast<hipfftDoubleComplex*>(out),
993-
HIPFFT_INVERSE));
993+
CHECK_CUFFT(hipfftExecZ2Z(plan, reinterpret_cast<hipfftDoubleComplex*>(in), reinterpret_cast<hipfftDoubleComplex*>(out),
994+
HIPFFT_BACKWARD));
994995
}
995996
#endif
996997

@@ -1059,7 +1060,6 @@ typename BatchedFFT<FPTYPE>::fftHandleType BatchedFFT<FPTYPE>::get_plan_from_cac
10591060
this->plans[batchSize] = plan;
10601061
return plan;
10611062
}
1062-
10631063
template class BatchedFFT<float>;
10641064
template class BatchedFFT<double>;
10651065

0 commit comments

Comments
 (0)