Neko 1.99.5
A portable framework for high-order spectral element flow simulations
Loading...
Searching...
No Matches
phmg.f90
Go to the documentation of this file.
1! Copyright (c) 2024-2026, The Neko Authors
2! All rights reserved.
3!
4! Redistribution and use in source and binary forms, with or without
5! modification, are permitted provided that the following conditions
6! are met:
7!
8! * Redistributions of source code must retain the above copyright
9! notice, this list of conditions and the following disclaimer.
10!
11! * Redistributions in binary form must reproduce the above
12! copyright notice, this list of conditions and the following
13! disclaimer in the documentation and/or other materials provided
14! with the distribution.
15!
16! * Neither the name of the authors nor the names of its
17! contributors may be used to endorse or promote products derived
18! from this software without specific prior written permission.
19!
20! THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
21! "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT
22! LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS
23! FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE
24! COPYRIGHT OWNER OR CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT,
25! INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,
26! BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;
27! LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER
28! CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT
29! LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN
30! ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF THE
31! POSSIBILITY OF SUCH DAMAGE.
32!
34module phmg
35 use num_types, only : rp
36 use precon, only : pc_t
37 use gather_scatter, only : gs_t, gs_op_add
38 use space, only : space_t, gll
39 use dofmap, only : dofmap_t
40 use field, only : field_t
41 use coefs, only : coef_t
42 use mesh, only : mesh_t
43 use bc, only : bc_t
44 use bc_list, only : bc_list_t
45 use dirichlet, only : dirichlet_t
46 use utils, only : neko_error
47 use cheby, only : cheby_t
49 use jacobi, only : jacobi_t
51 use schwarz, only : schwarz_t
52 use ax_product, only : ax_t, ax_helm_factory
55 use json_module, only : json_file
57 use math, only : copy, col2, add2, add2s2, add2s1
63 use krylov, only : ksp_t, ksp_monitor_t, ksp_max_iter, &
64 krylov_solver_factory
66 use logger, only : neko_log, log_size
67 use, intrinsic :: iso_c_binding
68 implicit none
69 private
70
71
72 type, private :: phmg_lvl_t
73 integer :: lvl = -1
74 integer :: smoother_itrs = 10
75 type(space_t), pointer :: xh
76 type(dofmap_t), pointer :: dm_xh
77 type(gs_t), pointer :: gs_h
79 type(cheby_t) :: cheby
83 type(coef_t), pointer :: coef
84 type(bc_list_t) :: bclst
85 type(dirichlet_t) :: bc
86 type(field_t) :: r, w, z
87 end type phmg_lvl_t
88
89 type, public :: phmg_hrchy_t
90 type(phmg_lvl_t), allocatable :: lvl(:)
91 end type phmg_hrchy_t
92
93
94 type, public, extends(pc_t) :: phmg_t
95 type(tamg_solver_t) :: amg_solver
96 integer :: nlvls
97 type(phmg_hrchy_t) :: phmg_hrchy
98 class(ax_t), allocatable :: ax
99 type(interpolator_t), allocatable :: intrp(:)
100 type(mesh_t), pointer :: msh
101 contains
102 procedure, pass(this) :: init => phmg_init
103 procedure, pass(this) :: init_from_components => &
105 procedure, pass(this) :: free => phmg_free
106 procedure, pass(this) :: solve => phmg_solve
107 procedure, pass(this) :: update => phmg_update
108 procedure, private, pass(this) :: mg_cycle => phmg_mg_cycle
109 end type phmg_t
110
111contains
112
113 subroutine phmg_init(this, coef, bclst, phmg_params)
114 class(phmg_t), intent(inout), target :: this
115 type(coef_t), intent(in), target :: coef
116 type(bc_list_t), intent(inout), target :: bclst
117 type(json_file), intent(inout) :: phmg_params
118 integer :: crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree
119 integer :: smoother_itrs
120 character(len=:), allocatable :: cheby_acc
121 integer, allocatable :: pcrs_sched(:)
122
123 call json_get_or_default(phmg_params, 'smoother_iterations', &
124 smoother_itrs, 3)
125
126 call json_get_or_default(phmg_params, 'smoother_cheby_acc', &
127 cheby_acc, "jacobi")
128
129 call json_get_or_default(phmg_params, 'coarse_grid.levels', &
130 crs_tamg_lvls, 3)
131
132 call json_get_or_default(phmg_params, 'coarse_grid.iterations', &
133 crs_tamg_itrs, 1)
134
135 call json_get_or_default(phmg_params, 'coarse_grid.cheby_degree', &
136 crs_tamg_cheby_degree, 4)
137
138 if (phmg_params%valid_path('pcoarsening_schedule')) then
139 call json_get(phmg_params, 'pcoarsening_schedule', pcrs_sched)
140 else
141 allocate(pcrs_sched(2))
142 pcrs_sched(1) = 3
143 pcrs_sched(2) = 1
144 end if
145
146
147 call this%init_from_components(coef, bclst, smoother_itrs, &
148 cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree,&
149 pcrs_sched)
150
151 end subroutine phmg_init
152
153 subroutine phmg_init_from_components(this, coef, bclst, smoother_itrs, &
154 cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree, &
155 pcrs_sched)
156 class(phmg_t), intent(inout), target :: this
157 type(coef_t), intent(in), target :: coef
158 type(bc_list_t), intent(inout), target :: bclst
159 integer, intent(in) :: smoother_itrs
160 character(len=:), allocatable :: cheby_acc
161 integer, intent(in) :: crs_tamg_lvls, crs_tamg_itrs
162 integer, intent(in) :: crs_tamg_cheby_degree
163 integer, intent(in), allocatable :: pcrs_sched(:)
164 integer :: lx_crs, lx_mid
165 integer, allocatable :: lx_lvls(:)
166 integer :: n, i, j, st
167 class(bc_t), pointer :: bc_j
168 logical :: use_jacobi, use_cheby
169 use_jacobi = .true.
170 use_cheby = .true.
171
172 this%msh => coef%msh
173
174 this%nlvls = size(pcrs_sched) + 1
175 allocate(lx_lvls(0:this%nlvls - 1))
176 lx_lvls(1:) = pcrs_sched + 1
177
178 allocate(this%phmg_hrchy%lvl(0:this%nlvls - 1))
179
180 this%phmg_hrchy%lvl(0)%lvl = 0
181 this%phmg_hrchy%lvl(0)%smoother_itrs = smoother_itrs
182 this%phmg_hrchy%lvl(0)%Xh => coef%Xh
183 this%phmg_hrchy%lvl(0)%coef => coef
184 this%phmg_hrchy%lvl(0)%dm_Xh => coef%dof
185 this%phmg_hrchy%lvl(0)%gs_h => coef%gs_h
186
187 do i = 1, this%nlvls - 1
188 allocate(this%phmg_hrchy%lvl(i)%Xh)
189 allocate(this%phmg_hrchy%lvl(i)%dm_Xh)
190 allocate(this%phmg_hrchy%lvl(i)%gs_h)
191 allocate(this%phmg_hrchy%lvl(i)%coef)
192
193 this%phmg_hrchy%lvl(i)%lvl = i
194 this%phmg_hrchy%lvl(i)%smoother_itrs = smoother_itrs
195 call this%phmg_hrchy%lvl(i)%Xh%init(gll, lx_lvls(i), lx_lvls(i), &
196 lx_lvls(i))
197 call this%phmg_hrchy%lvl(i)%dm_Xh%init(coef%msh, &
198 this%phmg_hrchy%lvl(i)%Xh)
199 call this%phmg_hrchy%lvl(i)%gs_h%init(this%phmg_hrchy%lvl(i)%dm_Xh)
200 call this%phmg_hrchy%lvl(i)%coef%init(this%phmg_hrchy%lvl(i)%gs_h)
201 end do
202
203 do i = 0, this%nlvls - 1
204 call this%phmg_hrchy%lvl(i)%r%init(this%phmg_hrchy%lvl(i)%dm_Xh)
205 call this%phmg_hrchy%lvl(i)%w%init(this%phmg_hrchy%lvl(i)%dm_Xh)
206 call this%phmg_hrchy%lvl(i)%z%init(this%phmg_hrchy%lvl(i)%dm_Xh)
207
208 this%phmg_hrchy%lvl(i)%coef%ifh2 = coef%ifh2
209 call copy(this%phmg_hrchy%lvl(i)%coef%h1, coef%h1, &
210 this%phmg_hrchy%lvl(i)%dm_Xh%size())
211
212 call this%phmg_hrchy%lvl(i)%bc%init_base(this%phmg_hrchy%lvl(i)%coef)
213 if (bclst%size() .gt. 0 ) then
214 do j = 1, bclst%size()
215 bc_j => bclst%get(j)
216 call this%phmg_hrchy%lvl(i)%bc%mark_facets(bc_j%marked_facet)
217 end do
218 end if
219 call this%phmg_hrchy%lvl(i)%bc%finalize()
220 call this%phmg_hrchy%lvl(i)%bc%set_g(0.0_rp)
221 call this%phmg_hrchy%lvl(i)%bclst%init()
222 call this%phmg_hrchy%lvl(i)%bclst%append(this%phmg_hrchy%lvl(i)%bc)
223
225 if (trim(cheby_acc) .eq. "schwarz") then
226 call this%phmg_hrchy%lvl(i)%schwarz%init( &
227 this%phmg_hrchy%lvl(i)%Xh, &
228 this%phmg_hrchy%lvl(i)%dm_Xh, &
229 this%phmg_hrchy%lvl(i)%gs_h, &
230 this%phmg_hrchy%lvl(i)%bclst, &
231 coef%msh)
232 end if
233
234 if (neko_bcknd_device .eq. 1) then
235 call this%phmg_hrchy%lvl(i)%device_jacobi%init(&
236 this%phmg_hrchy%lvl(i)%coef, &
237 this%phmg_hrchy%lvl(i)%dm_Xh, &
238 this%phmg_hrchy%lvl(i)%gs_h)
239 else
240 call this%phmg_hrchy%lvl(i)%jacobi%init(&
241 this%phmg_hrchy%lvl(i)%coef, &
242 this%phmg_hrchy%lvl(i)%dm_Xh, &
243 this%phmg_hrchy%lvl(i)%gs_h)
244 end if
245
246 if (neko_bcknd_device .eq. 1) then
247 if (trim(cheby_acc) .eq. "jacobi") then
248 call this%phmg_hrchy%lvl(i)%cheby_device%init( &
249 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs, &
250 this%phmg_hrchy%lvl(i)%device_jacobi)
251 st = 1
252 else
253 call this%phmg_hrchy%lvl(i)%cheby_device%init( &
254 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs)
255 st = 0
256 if (trim(cheby_acc) .eq. "schwarz") then
257 this%phmg_hrchy%lvl(i)%cheby_device%schwarz => &
258 this%phmg_hrchy%lvl(i)%schwarz
259 st = 2
260 end if
261 end if
262 else
263 if (trim(cheby_acc) .eq. "jacobi") then
264 call this%phmg_hrchy%lvl(i)%cheby%init( &
265 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs, &
266 this%phmg_hrchy%lvl(i)%jacobi)
267 st = 1
268 else
269 call this%phmg_hrchy%lvl(i)%cheby%init( &
270 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs)
271 st = 0
272 if (trim(cheby_acc) .eq. "schwarz") then
273 this%phmg_hrchy%lvl(i)%cheby%schwarz => &
274 this%phmg_hrchy%lvl(i)%schwarz
275 st = 2
276 end if
277 end if
278 end if
279
280 end do
281
282 call print_phmg_info(this%nlvls, st, this%phmg_hrchy)
283
284 ! Create backend specific Ax operator
285 call ax_helm_factory(this%ax, full_formulation = .false.)
286
287 ! Interpolator Fine + mg levels
288 allocate(this%intrp(this%nlvls - 1))
289 do i = 1, this%nlvls -1
290 call this%intrp(i)%init(this%phmg_hrchy%lvl(i-1)%Xh, &
291 this%phmg_hrchy%lvl(i)%Xh)
292 end do
293
294 call this%amg_solver%init(this%ax, this%phmg_hrchy%lvl(this%nlvls -1)%Xh, &
295 this%phmg_hrchy%lvl(this%nlvls -1)%coef, this%msh, &
296 this%phmg_hrchy%lvl(this%nlvls-1)%gs_h, crs_tamg_lvls, &
297 this%phmg_hrchy%lvl(this%nlvls -1)%bclst, &
298 crs_tamg_itrs, crs_tamg_cheby_degree)
299
300 end subroutine phmg_init_from_components
301
302 subroutine phmg_free(this)
303 class(phmg_t), intent(inout) :: this
304 end subroutine phmg_free
305
306 subroutine phmg_solve(this, z, r, n)
307 class(phmg_t), intent(inout) :: this
308 integer, intent(in) :: n
309 real(kind=rp), dimension(n), intent(inout) :: z
310 real(kind=rp), dimension(n), intent(inout) :: r
311 type(c_ptr) :: z_d, r_d
312 type(ksp_monitor_t) :: ksp_results
313 integer :: i
314
315 call profiler_start_region('PHMG_solve', 8)
316 associate( mglvl => this%phmg_hrchy%lvl)
317 if (neko_bcknd_device .eq. 1) then
318 z_d = device_get_ptr(z)
319 r_d = device_get_ptr(r)
320 !We should not work with the input
321 call device_copy(mglvl(0)%r%x_d, r_d, n)
322
323 call device_rzero(mglvl(0)%z%x_d, n)
324 call device_rzero(mglvl(0)%w%x_d, n)
325
326 call this%mg_cycle()
327
328 call device_copy(z_d, mglvl(0)%z%x_d, n)
329 else
330 !OCL NORECURRENCE, NOVREC, NOALIAS
331 !DIR$ CONCURRENT
332 !GCC$ ivdep
333 !$omp parallel do
334 do i = 1, n
335 !We should not work with the input
336 mglvl(0)%r%x(i,1,1,1) = r(i)
337
338 mglvl(0)%z%x(i,1,1,1) = 0.0_rp
339 mglvl(0)%w%x(i,1,1,1) = 0.0_rp
340 end do
341 !$omp end parallel do
342
343 call this%mg_cycle()
344
345 !OCL NORECURRENCE, NOVREC, NOALIAS
346 !DIR$ CONCURRENT
347 !GCC$ ivdep
348 !$omp parallel do
349 do i = 1, n
350 z(i) = mglvl(0)%z%x(i,1,1,1)
351 end do
352 !$omp end parallel do
353 end if
354 end associate
355 call profiler_end_region('PHMG_solve', 8)
356
357 end subroutine phmg_solve
358
359 subroutine phmg_update(this)
360 class(phmg_t), intent(inout) :: this
361 end subroutine phmg_update
362
363
364 subroutine phmg_mg_cycle(this)
365 class(phmg_t), intent(inout) :: this
366 type(ksp_monitor_t) :: ksp_results
367 character(len=2) :: lvl_name
368 integer :: lvl, i
369
370 associate(mg => this%phmg_hrchy%lvl, intrp => this%intrp, &
371 msh => this%msh, ax => this%Ax)
372 do lvl = 0, this%nlvls-2
373 write(lvl_name, '(I0)') lvl
374 call profiler_start_region( "PHMG_level_" // trim(lvl_name))
375 associate(z => mg(lvl)%z, r => mg(lvl)%r, w => mg(lvl)%w)
376 !------------!
377 ! SMOOTH !
378 !------------!
379 if (neko_bcknd_device .eq. 1) then
380 mg(lvl)%cheby_device%zero_initial_guess = .true.
381 ksp_results = mg(lvl)%cheby_device%solve(ax, z, &
382 r%x, mg(lvl)%dm_Xh%size(), &
383 mg(lvl)%coef, mg(lvl)%bclst, &
384 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
385 else
386 mg(lvl)%cheby%zero_initial_guess = .true.
387 ksp_results = mg(lvl)%cheby%solve(ax, z, &
388 r%x, mg(lvl)%dm_Xh%size(), &
389 mg(lvl)%coef, mg(lvl)%bclst, &
390 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
391 end if
392
393 !------------!
394 ! Residual !
395 !------------!
396 call ax%compute(w%x, z%x, mg(lvl)%coef, msh, mg(lvl)%Xh)
397 call mg(lvl)%gs_h%op(w%x, mg(lvl)%dm_Xh%size(), gs_op_add, &
398 glb_cmd_event)
399 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
400 call mg(lvl)%bclst%apply_scalar(w%x, mg(lvl)%dm_Xh%size())
401
402 if (neko_bcknd_device .eq. 1) then
403 call device_add2s1(w%x_d, r%x_d, -1.0_rp, mg(lvl)%dm_Xh%size())
404 else
405 !OCL NORECURRENCE, NOVREC, NOALIAS
406 !DIR$ CONCURRENT
407 !GCC$ ivdep
408 !$omp parallel do
409 do i = 1, mg(lvl)%dm_Xh%size()
410 w%x(i,1,1,1) = r%x(i,1,1,1) - w%x(i,1,1,1)
411 end do
412 !$omp end parallel do
413 end if
414
415 !------------!
416 ! Restrict !
417 !------------!
418 if (neko_bcknd_device .eq. 1) then
419 call device_col2(w%x_d, mg(lvl)%coef%mult_d, mg(lvl)%dm_Xh%size())
420 else
421 call col2(w%x, mg(lvl)%coef%mult, mg(lvl)%dm_Xh%size())
422 end if
423
424 call intrp(lvl+1)%map(mg(lvl+1)%r%x, w%x, msh%nelv, mg(lvl+1)%Xh)
425
426 call mg(lvl+1)%gs_h%op(mg(lvl+1)%r%x, mg(lvl+1)%dm_Xh%size(), &
427 gs_op_add, glb_cmd_event)
428 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
429
430 call mg(lvl+1)%bclst%apply_scalar( &
431 mg(lvl+1)%r%x, &
432 mg(lvl+1)%dm_Xh%size())
433
434 if (neko_bcknd_device .eq. 1) then
435 call device_rzero(mg(lvl+1)%z%x_d, mg(lvl+1)%dm_Xh%size())
436 else
437 !OCL NORECURRENCE, NOVREC, NOALIAS
438 !DIR$ CONCURRENT
439 !GCC$ ivdep
440 !$omp parallel do
441 do i = 1, mg(lvl+1)%dm_Xh%size()
442 mg(lvl+1)%z%x(i,1,1,1) = 0.0_rp
443 end do
444 !$omp end parallel do
445 end if
446 end associate
447 call profiler_end_region( "PHMG_level_" // trim(lvl_name))
448 end do
449
450 call profiler_start_region( 'PHMG_coarse-solve' )
451 !------------!
452 ! SOLVE !
453 !------------!
454 call this%amg_solver%solve(mg(this%nlvls-1)%z%x, &
455 mg(this%nlvls-1)%r%x, &
456 mg(this%nlvls-1)%dm_Xh%size())
457 call profiler_end_region( 'PHMG_coarse-solve' )
458
459 do lvl = (this%nlvls-2), 0, -1
460 write(lvl_name, '(I0)') lvl
461 call profiler_start_region( "PHMG_level_" // trim(lvl_name))
462 associate(z => mg(lvl)%z, r => mg(lvl)%r, w => mg(lvl)%w)
463 !------------!
464 ! Project !
465 !------------!
466 call intrp(lvl+1)%map(w%x, mg(lvl+1)%z%x, msh%nelv, mg(lvl)%Xh)
467
468 call mg(lvl)%gs_h%op(w%x, mg(lvl)%dm_Xh%size(), gs_op_add, &
469 glb_cmd_event)
470 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
471
472 if (neko_bcknd_device .eq. 1) then
473 call device_col2(w%x_d, mg(lvl)%coef%mult_d, mg(lvl)%dm_Xh%size())
474 else
475 call col2(w%x, mg(lvl)%coef%mult, mg(lvl)%dm_Xh%size())
476 end if
477
478 !------------!
479 ! Correct !
480 !------------!
481 if (neko_bcknd_device .eq. 1) then
482 call device_add2(z%x_d, w%x_d, mg(lvl)%dm_Xh%size())
483 else
484 !OCL NORECURRENCE, NOVREC, NOALIAS
485 !DIR$ CONCURRENT
486 !GCC$ ivdep
487 !$omp parallel do
488 do i = 1, mg(lvl)%dm_Xh%size()
489 z%x(i,1,1,1) = z%x(i,1,1,1) + w%x(i,1,1,1)
490 end do
491 !$omp end parallel do
492 end if
493
494 !------------!
495 ! SMOOTH !
496 !------------!
497 if (neko_bcknd_device .eq. 1) then
498 ksp_results = mg(lvl)%cheby_device%solve(ax, z, &
499 r%x, mg(lvl)%dm_Xh%size(), &
500 mg(lvl)%coef, mg(lvl)%bclst, &
501 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
502 else
503 ksp_results = mg(lvl)%cheby%solve(ax, z, &
504 r%x, mg(lvl)%dm_Xh%size(), &
505 mg(lvl)%coef, mg(lvl)%bclst, &
506 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
507 end if
508 end associate
509 call profiler_end_region( "PHMG_level_" // trim(lvl_name))
510 end do
511 end associate
512
513 end subroutine phmg_mg_cycle
514
524 subroutine phmg_jacobi_smoother(z, r, w, mg, msh, Ax, n, lvl)
525 type(phmg_lvl_t) :: mg
526 class(ax_t), intent(inout) :: Ax
527 type(mesh_t), intent(inout) :: msh
528 type(field_t), intent(inout) :: z, r, w
529 integer, intent(in) :: n, lvl
530 integer :: i, j, iblk, ni, niblk
531
532 ni = mg%smoother_itrs
533 if (neko_bcknd_device .eq. 1) then
534 do i = 1, ni
535 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
536 call mg%gs_h%op(w%x, n, gs_op_add, glb_cmd_event)
537 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
538 call mg%bclst%apply_scalar(w%x, n)
539 call device_add2s1(w%x_d, r%x_d, -1.0_rp, n)
540
541 call mg%device_jacobi%solve(w%x, w%x, n)
542
543 call device_add2s2(z%x_d, w%x_d, 0.6_rp, n)
544 end do
545 else
546 do i = 1, ni
547 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
548 call mg%gs_h%op(w%x, n, gs_op_add)
549 call mg%bclst%apply_scalar(w%x, n)
550 call add2s1(w%x, r%x, -1.0_rp, n)
551
552 call mg%jacobi%solve(w%x, w%x, n)
553
554 call add2s2(z%x, w%x, 0.6_rp, n)
555 end do
556 end if
557 end subroutine phmg_jacobi_smoother
558
559
560 subroutine phmg_resid_monitor(z, r, w, mg, msh, Ax, lvl, typ)
561 integer :: lvl, typ
562 type(phmg_lvl_t) :: mg
563 class(ax_t), intent(inout) :: Ax
564 type(mesh_t), intent(inout) :: msh
565 type(field_t) :: z, r, w
566 real(kind=rp) :: val
567 character(len=LOG_SIZE) :: log_buf
568 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
569 call mg%gs_h%op(w%x, mg%dm_Xh%size(), gs_op_add)
570 call mg%bclst%apply_scalar(w%x, mg%dm_Xh%size())
571 call device_add2s1(w%x_d, r%x_d, -1.0_rp, mg%dm_Xh%size())
572 val = device_glsc2(w%x_d, w%x_d, mg%dm_Xh%size())
573 if (typ .eq. 1) then
574 write(log_buf, '(A15,I4,F12.6)') 'PRESMOO - PRE', lvl, val
575 else if (typ .eq. 2) then
576 write(log_buf, '(A15,I4,F12.6)') 'PRESMOO -POST', lvl, val
577 else if (typ .eq. 3) then
578 write(log_buf, '(A15,I4,F12.6)') 'POSTSMOO- PRE', lvl, val
579 else if (typ .eq. 4) then
580 write(log_buf, '(A15,I4,F12.6)') 'POSTSMOO-POST', lvl, val
581 else if (typ .eq. 5) then
582 write(log_buf, '(A15,I4,F12.6)') 'TAMG - PRE', lvl, val
583 else if (typ .eq. 6) then
584 write(log_buf, '(A15,I4,F12.6)') 'TAMG -POST', lvl, val
585 else
586 write(log_buf, '(A15,I4,F12.6)') 'RESID', lvl, val
587 end if
588 call neko_log%message(log_buf)
589 end subroutine phmg_resid_monitor
590
591 subroutine print_phmg_info(nlvls, smoo_type, phmg)
592 integer, intent(in) :: nlvls
593 integer, intent(in) :: smoo_type
594 type(phmg_hrchy_t) :: phmg
595 integer :: i, clvl
596 character(len=LOG_SIZE) :: log_buf, smoo_name
597
598 call neko_log%section('PHMG')
599
600 if (smoo_type .eq. 1) then
601 write(smoo_name, '(A16)') 'CHEBY-acc JACOBI'
602 else if (smoo_type .eq. 2) then
603 write(smoo_name, '(A17)') 'CHEBY-acc SCHWARZ'
604 else
605 write(smoo_name, '(A5)') 'CHEBY'
606 end if
607
608 write(log_buf, '(A28,I2,A8)') &
609 'Creating PHMG hierarchy with', &
610 nlvls, 'levels.'
611 call neko_log%message(log_buf)
612
613 clvl = nlvls - 1
614 do i = 0, nlvls-1
615 write(log_buf, '(A8,I2,A8,I2)') &
616 '-- level', i, '-- lx:', phmg%lvl(i)%Xh%lx
617 call neko_log%message(log_buf)
618
619 if (i .eq. clvl) then
620 write(log_buf, '(A19,A20)') &
621 'Solve:', 'tAMG'
622 call neko_log%message(log_buf)
623 else
624 write(log_buf, '(A22,A20)') &
625 'Smoother:', &
626 trim(smoo_name)
627 call neko_log%message(log_buf)
628
629 write(log_buf, '(A28,I2)') &
630 'Smoother Iters:', &
631 phmg%lvl(i)%smoother_itrs
632 call neko_log%message(log_buf)
633 end if
634 end do
635
636 call neko_log%end_section()
637
638 end subroutine print_phmg_info
639
640end module phmg
__device__ T solve(const T u, const T y, const T guess, const T nu, const T kappa, const T B)
Return the device pointer for an associated Fortran array.
Definition device.F90:108
Retrieves a parameter by name or assigns a provided default value. In the latter case also adds the m...
Retrieves a parameter by name or throws an error.
Defines a Matrix-vector product.
Definition ax.f90:34
Defines a list of bc_t.
Definition bc_list.f90:34
Defines a boundary condition.
Definition bc.f90:34
Chebyshev preconditioner.
Chebyshev preconditioner.
Definition cheby.f90:34
Coefficients.
Definition coef.f90:34
Jacobi preconditioner accelerator backend.
subroutine, public device_add2s1(a_d, b_d, c1, n, strm)
subroutine, public device_add2s2(a_d, b_d, c1, n, strm)
Vector addition with scalar multiplication (multiplication on first argument)
subroutine, public device_add2(a_d, b_d, n, strm)
Vector addition .
subroutine, public device_rzero(a_d, n, strm)
Zero a real vector.
subroutine, public device_copy(a_d, b_d, n, strm)
Copy a vector .
subroutine, public device_col2(a_d, b_d, n, strm)
Vector multiplication .
subroutine, public device_invcol2(a_d, b_d, n, strm)
Vector division .
real(kind=rp) function, public device_glsc2(a_d, b_d, n, strm)
Weighted inner product .
Device abstraction, common interface for various accelerators.
Definition device.F90:34
subroutine, public device_stream_wait_event(stream, event, flags)
Synchronize a device stream with an event.
Definition device.F90:1471
type(c_ptr), bind(C), public glb_cmd_queue
Global command queue.
Definition device.F90:52
type(c_ptr), bind(C), public glb_cmd_event
Event for the global command queue.
Definition device.F90:63
Defines a dirichlet boundary condition.
Definition dirichlet.f90:34
Defines a mapping of the degrees of freedom.
Definition dofmap.f90:35
Defines a field.
Definition field.f90:34
Gather-scatter.
Routines to interpolate between different spaces.
Jacobi preconditioner.
Definition pc_jacobi.f90:34
Utilities for retrieving parameters from the case files.
Implements the base abstract type for Krylov solvers plus helper types.
Definition krylov.f90:34
integer, parameter, public ksp_max_iter
Maximum number of iters.
Definition krylov.f90:51
Logging routines.
Definition log.f90:34
type(log_t), public neko_log
Global log stream.
Definition log.f90:80
integer, parameter, public log_size
Definition log.f90:46
Definition math.f90:60
subroutine, public add2s1(a, b, c1, n)
Vector addition with scalar multiplication (multiplication on first argument)
Definition math.f90:981
subroutine, public add2(a, b, n)
Vector addition .
Definition math.f90:900
subroutine, public col2(a, b, n)
Vector multiplication .
Definition math.f90:1046
subroutine, public copy(a, b, n)
Copy a vector .
Definition math.f90:291
subroutine, public add2s2(a, b, c1, n)
Vector addition with scalar multiplication (multiplication on second argument)
Definition math.f90:998
Defines a mesh.
Definition mesh.f90:34
Build configurations.
integer, parameter neko_bcknd_device
integer, parameter, public rp
Global precision used in computations.
Definition num_types.f90:12
Hybrid ph-multigrid preconditioner.
Definition phmg.f90:34
subroutine phmg_jacobi_smoother(z, r, w, mg, msh, ax, n, lvl)
Wraps jacobi solve as a residual update relaxation method.
Definition phmg.f90:525
subroutine phmg_init_from_components(this, coef, bclst, smoother_itrs, cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree, pcrs_sched)
Definition phmg.f90:156
subroutine phmg_solve(this, z, r, n)
Definition phmg.f90:307
subroutine phmg_init(this, coef, bclst, phmg_params)
Definition phmg.f90:114
subroutine phmg_update(this)
Definition phmg.f90:360
subroutine phmg_free(this)
Definition phmg.f90:303
subroutine phmg_mg_cycle(this)
Definition phmg.f90:365
subroutine phmg_resid_monitor(z, r, w, mg, msh, ax, lvl, typ)
Definition phmg.f90:561
subroutine print_phmg_info(nlvls, smoo_type, phmg)
Definition phmg.f90:592
Krylov preconditioner.
Definition precon.f90:34
Profiling interface.
Definition profiler.F90:34
subroutine, public profiler_start_region(name, region_id)
Started a named (name) profiler region.
Definition profiler.F90:79
subroutine, public profiler_end_region(name, region_id)
End the most recently started profiler region.
Definition profiler.F90:116
Overlapping schwarz solves.
Definition schwarz.f90:61
Defines a function space.
Definition space.f90:34
integer, parameter, public gll
Definition space.f90:49
Implements multigrid using the TreeAMG hierarchy structure. USE:
Utilities.
Definition utils.f90:35
Base type for a matrix-vector product providing .
Definition ax.f90:43
Base type for a boundary condition.
Definition bc.f90:62
A list of allocatable `bc_t`. Follows the standard interface of lists.
Definition bc_list.f90:49
Defines a Chebyshev preconditioner.
Definition cheby.f90:52
Defines a Chebyshev preconditioner.
Coefficients defined on a given (mesh, ) tuple. Arrays use indices (i,j,k,e): element e,...
Definition coef.f90:62
Defines a jacobi preconditioner.
Generic Dirichlet boundary condition on .
Definition dirichlet.f90:49
Interpolation between two space::space_t.
Defines a jacobi preconditioner.
Definition pc_jacobi.f90:45
Type for storing initial and final residuals in a Krylov solver.
Definition krylov.f90:56
Base abstract type for a canonical Krylov method, solving .
Definition krylov.f90:73
Defines a canonical Krylov preconditioner.
Definition precon.f90:40
The function space for the SEM solution fields.
Definition space.f90:63
Type for the TreeAMG solver.