Neko 1.99.6
A portable framework for high-order spectral element flow simulations
Loading...
Searching...
No Matches
pc_hsmg.f90
Go to the documentation of this file.
1! Copyright (c) 2008-2020, UCHICAGO ARGONNE, LLC.
2!
3! The UChicago Argonne, LLC as Operator of Argonne National
4! Laboratory holds copyright in the Software. The copyright holder
5! reserves all rights except those expressly granted to licensees,
6! and U.S. Government license rights.
7!
8! Redistribution and use in source and binary forms, with or without
9! modification, are permitted provided that the following conditions
10! are met:
11!
12! 1. Redistributions of source code must retain the above copyright
13! notice, this list of conditions and the disclaimer below.
14!
15! 2. Redistributions in binary form must reproduce the above copyright
16! notice, this list of conditions and the disclaimer (as noted below)
17! in the documentation and/or other materials provided with the
18! distribution.
19!
20! 3. Neither the name of ANL nor the names of its contributors
21! may be used to endorse or promote products derived from this software
22! without specific prior written permission.
23!
24! THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
25! "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT
26! LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS
27! FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL
28! UCHICAGO ARGONNE, LLC, THE U.S. DEPARTMENT OF
29! ENERGY OR CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL,
30! SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED
31! TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE,
32! DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY
33! THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
34! (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
35! OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
36!
37! Additional BSD Notice
38! ---------------------
39! 1. This notice is required to be provided under our contract with
40! the U.S. Department of Energy (DOE). This work was produced at
41! Argonne National Laboratory under Contract
42! No. DE-AC02-06CH11357 with the DOE.
43!
44! 2. Neither the United States Government nor UCHICAGO ARGONNE,
45! LLC nor any of their employees, makes any warranty,
46! express or implied, or assumes any liability or responsibility for the
47! accuracy, completeness, or usefulness of any information, apparatus,
48! product, or process disclosed, or represents that its use would not
49! infringe privately-owned rights.
50!
51! 3. Also, reference herein to any specific commercial products, process,
52! or services by trade name, trademark, manufacturer or otherwise does
53! not necessarily constitute or imply its endorsement, recommendation,
54! or favoring by the United States Government or UCHICAGO ARGONNE LLC.
55! The views and opinions of authors expressed
56! herein do not necessarily state or reflect those of the United States
57! Government or UCHICAGO ARGONNE, LLC, and shall
58! not be used for advertising or product endorsement purposes.
59!
61module hsmg
63 use num_types, only : rp
64 use math, only : copy, col2, add2
65 use utils, only : neko_error
66 use precon, only : pc_t, precon_factory, precon_destroy
67 use ax_product, only : ax_t, ax_helm_factory
68 use gather_scatter, only : gs_t, gs_op_add
70 use bc, only : bc_t
71 use bc_list, only : bc_list_t
72 use dirichlet, only : dirichlet_t
73 use schwarz, only : schwarz_t
74 use jacobi, only : jacobi_t
75 use sx_jacobi, only : sx_jacobi_t
81 use space, only : space_t, gll
82 use dofmap, only : dofmap_t
83 use field, only : field_t
84 use coefs, only : coef_t
85 use mesh, only : mesh_t
86 use json_module, only : json_file
88 use krylov, only : ksp_t, ksp_monitor_t, ksp_max_iter, &
89 krylov_solver_factory
92 use logger, only : neko_log, log_size
93 use, intrinsic :: iso_c_binding, only : c_ptr, c_null_ptr, c_associated
94 !$ use omp_lib
95 implicit none
96 private
97
98 !Struct to arrange our multigridlevels
99 type, private :: multigrid_t
100 type(dofmap_t), pointer :: dof => null()
101 type(gs_t), pointer :: gs_h => null()
102 type(space_t), pointer :: xh => null()
103 type(coef_t), pointer :: coef => null()
104 type(bc_list_t), pointer :: bclst => null()
105 type(schwarz_t), pointer :: schwarz => null()
106 type(field_t), pointer :: e => null()
107 end type multigrid_t
108
109 type, public, extends(pc_t) :: hsmg_t
110 type(mesh_t), pointer :: msh => null()
111 integer :: nlvls
112 type(multigrid_t), allocatable :: grids(:)
113 type(gs_t) :: gs_crs, gs_mg
114 type(space_t) :: xh_crs, xh_mg
115 type(dofmap_t) :: dm_crs, dm_mg
116 type(coef_t) :: c_crs, c_mg
117 type(zero_dirichlet_t) :: bc_crs, bc_mg, bc_reg
118 type(bc_list_t) :: bclst_crs, bclst_mg, bclst_reg
119 type(schwarz_t) :: schwarz, schwarz_mg, schwarz_crs
121 type(field_t) :: e, e_mg, e_crs
122 type(field_t) :: wf
123 class(ksp_t), allocatable :: crs_solver
124 type(tamg_solver_t), allocatable :: amg_solver
125 integer :: niter
126 class(pc_t), allocatable :: pc_crs
127 class(ax_t), allocatable :: ax
128 real(kind=rp), allocatable :: r(:)
129 type(interpolator_t) :: interp_fine_mid
130 type(interpolator_t) :: interp_mid_crs
131 real(kind=rp), allocatable :: w(:)
132 type(c_ptr) :: w_d = c_null_ptr
133 type(c_ptr) :: r_d = c_null_ptr
134 type(c_ptr) :: hsmg_event = c_null_ptr
135 type(c_ptr) :: gs_event = c_null_ptr
136 contains
137 procedure, pass(this) :: init => hsmg_init
138 procedure, pass(this) :: init_from_components => &
140 procedure, pass(this) :: free => hsmg_free
141 procedure, pass(this) :: solve => hsmg_solve
142 procedure, pass(this) :: update => hsmg_set_h
143 end type hsmg_t
144
145contains
146
147 subroutine hsmg_init(this, coef, bclst, hsmg_params)
148 class(hsmg_t), intent(inout), target :: this
149 type(coef_t), intent(in), target :: coef
150 type(bc_list_t), intent(inout), target :: bclst
151 type(json_file), intent(inout) :: hsmg_params
152 character(len=:), allocatable :: crs_solver, crs_pc
153 logical :: crs_monitor
154 integer :: crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree
155
156 ! Exract coarse grid parameters
157
158 ! Common parameters for the coarse grid
159 call json_get_or_default(hsmg_params, 'coarse_grid.solver', &
160 crs_solver, "cg")
161
162 !
163 ! Parameters for a Krylov based coarse grid solverthis
164 !
165 call json_get_or_default(hsmg_params, 'coarse_grid.iterations', &
166 this%niter, 10)
167
168 call json_get_or_default(hsmg_params, 'coarse_grid.preconditioner', &
169 crs_pc, "jacobi")
170
171 call json_get_or_default(hsmg_params, 'coarse_grid.monitor', &
172 crs_monitor, .false.)
173
174 !
175 ! Parameters for a tree-amg based coarse grid solver
176 !
177 call json_get_or_default(hsmg_params, 'coarse_grid.levels', &
178 crs_tamg_lvls, 3)
179
180 call json_get_or_default(hsmg_params, 'coarse_grid.iterations', &
181 crs_tamg_itrs, 1)
182
183 call json_get_or_default(hsmg_params, 'coarse_grid.cheby_degree', &
184 crs_tamg_cheby_degree, 4)
185
186 call this%init_from_components(coef, bclst, crs_solver, crs_pc, &
187 crs_monitor, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree)
188
189 end subroutine hsmg_init
190
191 subroutine hsmg_init_from_components(this, coef, bclst, crs_solver, crs_pc, &
192 crs_monitor, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree)
193 class(hsmg_t), intent(inout), target :: this
194 type(coef_t), intent(in), target :: coef
195 type(bc_list_t), intent(inout), target :: bclst
196 character(len=:), intent(inout), allocatable :: crs_solver, crs_pc
197 logical, intent(inout) :: crs_monitor
198 integer, intent(in) :: crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree
199 integer :: n, i
200 integer :: lx_crs, lx_mid
201 class(bc_t), pointer :: bc_i
202
203 character(len=LOG_SIZE) :: log_buf
204
205 call this%free()
208 this%nlvls = 3
209 lx_crs = 2
210 if (coef%Xh%lx .lt. 5) then
211 lx_mid = max(coef%Xh%lx-1,3)
212
213 if (coef%Xh%lx .le. 2) then
214 call neko_error('Polynomial order < 2 not supported for hsmg precon')
215 end if
216
217 else
218 lx_mid = 4
219 end if
220
221
222 call neko_log%section('HSMG')
223 if (this%nlvls .lt. 1e1) then
224 write(log_buf, '(A,I1,A)') 'HSMG hierarchy : ', &
225 this%nlvls, ' levels'
226 else if (this%nlvls .lt. 1e2) then
227 write(log_buf, '(A,I2,A)') 'HSMG hierarchy : ', &
228 this%nlvls, ' levels'
229 else if (this%nlvls .lt. 1e3) then
230 write(log_buf, '(A,I3,A)') 'HSMG hierarchy : ', this%nlvls, &
231 ' levels'
232 else
233 write(log_buf, '(A,I6,A)') 'HSMG hierarchy : ', this%nlvls, &
234 ' levels'
235 end if
236 call neko_log%message(log_buf)
237 if (trim(crs_solver) .ne. 'tamg' .or. trim(crs_solver) .eq. 'cheby') then
238 call neko_log%message('Coarse grid solver : (' // trim(crs_solver) // &
239 ', ' // trim(crs_pc) // ')')
240
241 if (this%niter .lt. 1e1) then
242 write(log_buf, '(A,I1)') 'Coarse grid iters. : ', this%niter
243 else if (this%niter .lt. 1e2) then
244 write(log_buf, '(A,I2)') 'Coarse grid iters. : ', this%niter
245 else if (this%niter .lt. 1e3) then
246 write(log_buf, '(A,I3)') 'Coarse grid iters. : ', this%niter
247 else if (this%niter .lt. 1e4) then
248 write(log_buf, '(A,I4)') 'Coarse grid iters. : ', this%niter
249 else
250 write(log_buf, '(A,I6)') 'Coarse grid iters. : ', this%niter
251 end if
252
253 call neko_log%message(log_buf)
254 else
255 call neko_log%message('Coarse grid solver : ' // trim(crs_solver) )
256 end if
257
258 this%msh => coef%msh
259 allocate(this%grids(this%nlvls))
260 allocate(this%w(coef%dof%size()))
261 allocate(this%r(coef%dof%size()))
262
263
264 ! Compute all elements as if they are deformed
265 call coef%msh%all_deformed()
266
267 n = coef%dof%size()
268 call this%e%init(coef%dof, 'work array')
269 call this%wf%init(coef%dof, 'work 2')
270
271 call this%Xh_crs%init(gll, lx_crs, lx_crs, lx_crs)
272 call this%dm_crs%init(coef%msh, this%Xh_crs)
273 call this%gs_crs%init(this%dm_crs)
274 call this%e_crs%init(this%dm_crs, 'work crs')
275 call this%c_crs%init(this%gs_crs)
276
277 call this%Xh_mg%init(gll, lx_mid, lx_mid, lx_mid)
278 call this%dm_mg%init(coef%msh, this%Xh_mg)
279 call this%gs_mg%init(this%dm_mg)
280 call this%e_mg%init(this%dm_mg, 'work midl')
281 call this%c_mg%init(this%gs_mg)
282
283 ! Create backend specific Ax operator
284 call ax_helm_factory(this%ax, full_formulation = .false.)
285
286 call this%bc_crs%init_base(this%c_crs)
287 call this%bc_mg%init_base(this%c_mg)
288 call this%bc_reg%init_base(coef)
289 if (bclst%size() .gt. 0) then
290 do i = 1, bclst%size()
291 bc_i => bclst%get(i)
292 call this%bc_reg%mark_facets(bc_i%marked_facet)
293 bc_i => bclst%get(i)
294 call this%bc_crs%mark_facets(bc_i%marked_facet)
295 bc_i => bclst%get(i)
296 call this%bc_mg%mark_facets(bc_i%marked_facet)
297 end do
298 end if
299 call this%bc_reg%finalize()
300 call this%bc_crs%finalize()
301 call this%bc_mg%finalize()
302
303 call this%bclst_reg%init()
304 call this%bclst_crs%init()
305 call this%bclst_mg%init()
306
307 call this%bclst_reg%append(this%bc_reg)
308 call this%bclst_crs%append(this%bc_crs)
309 call this%bclst_mg%append(this%bc_mg)
310
311 call this%schwarz%init(coef%Xh, coef%dof, coef%gs_h, &
312 this%bclst_reg, coef%msh)
313 call this%schwarz_mg%init(this%Xh_mg, this%dm_mg, this%gs_mg,&
314 this%bclst_mg, coef%msh)
315
316 call this%interp_fine_mid%init(coef%Xh, this%Xh_mg)
317 call this%interp_mid_crs%init(this%Xh_mg, this%Xh_crs)
318
319 call hsmg_fill_grid(coef%dof, coef%gs_h, coef%Xh, coef, &
320 this%bclst_reg, this%schwarz, this%e, this%grids, 3)
321 call hsmg_fill_grid(this%dm_mg, this%gs_mg, this%Xh_mg, this%c_mg, &
322 this%bclst_mg, this%schwarz_mg, this%e_mg, &
323 this%grids, 2)
324 call hsmg_fill_grid(this%dm_crs, this%gs_crs, this%Xh_crs, &
325 this%c_crs, this%bclst_crs, this%schwarz_crs, &
326 this%e_crs, this%grids, 1)
327
328 call hsmg_set_h(this)
329 if (neko_bcknd_device .eq. 1) then
330 call device_map(this%w, this%w_d, n)
331 call device_map(this%r, this%r_d, n)
332 end if
333
334 call device_event_create(this%hsmg_event, 2)
335 call device_event_create(this%gs_event, 2)
336
337
338
339 ! Create a backend specific krylov solver
340 if (trim(crs_solver) .eq. 'tamg') then
341 allocate(this%amg_solver)
342 call this%amg_solver%init(this%ax, this%grids(1)%e%Xh, &
343 this%grids(1)%coef, this%msh, this%grids(1)%gs_h, crs_tamg_lvls, &
344 this%grids(1)%bclst, crs_tamg_itrs, crs_tamg_cheby_degree)
345 else
346 ! Create a backend specific preconditioner
347 call precon_factory(this%pc_crs, crs_pc)
348
349 select type (pc => this%pc_crs)
350 type is (jacobi_t)
351 call pc%init(this%c_crs, this%dm_crs, this%gs_crs)
352 type is (sx_jacobi_t)
353 call pc%init(this%c_crs, this%dm_crs, this%gs_crs)
354 type is (device_jacobi_t)
355 call pc%init(this%c_crs, this%dm_crs, this%gs_crs)
356 end select
357
358 call krylov_solver_factory(this%crs_solver, &
359 this%dm_crs%size(), trim(crs_solver), ksp_max_iter, &
360 m = this%pc_crs, monitor = crs_monitor)
361 end if
362
363 call neko_log%end_section()
364
365 end subroutine hsmg_init_from_components
366
367 subroutine hsmg_set_h(this)
368 class(hsmg_t), intent(inout) :: this
369 ! integer :: i
370 ! Yeah I dont really know what to do here. For incompressible flow not
371 ! much happens
372 this%grids(1)%coef%ifh2 = .false.
373 call copy(this%grids(1)%coef%h1, this%grids(3)%coef%h1, &
374 this%grids(1)%dof%size())
375 if (neko_bcknd_device .eq. 1) then
376 call device_copy(this%grids(1)%coef%h1_d, this%grids(3)%coef%h1_d, &
377 this%grids(1)%dof%size())
378 end if
379 end subroutine hsmg_set_h
380
381
382 subroutine hsmg_fill_grid(dof, gs_h, Xh, coef, bclst, schwarz, e, grids, l)
383 type(dofmap_t), target, intent(in) :: dof
384 type(gs_t), target, intent(in) :: gs_h
385 type(space_t), target, intent(in) :: Xh
386 type(coef_t), target, intent(in) :: coef
387 type(bc_list_t), target, intent(in) :: bclst
388 type(schwarz_t), target, intent(in) :: schwarz
389 type(field_t), target, intent(in) :: e
390 integer, intent(in) :: l
391 type(multigrid_t), intent(inout), dimension(l) :: grids
392
393
394 grids(l)%dof => dof
395 grids(l)%gs_h => gs_h
396 grids(l)%Xh => xh
397 grids(l)%coef => coef
398 grids(l)%bclst => bclst
399 grids(l)%schwarz => schwarz
400 grids(l)%e => e
401
402 end subroutine hsmg_fill_grid
403
404 subroutine hsmg_free(this)
405 class(hsmg_t), intent(inout) :: this
406
407 if (allocated(this%ax)) then
408 deallocate(this%ax)
409 end if
410
411 if (allocated(this%grids)) then
412 deallocate(this%grids)
413 end if
414
415 if (allocated(this%w)) then
416 if (c_associated(this%w_d)) then
417 call device_unmap(this%w, this%w_d)
418 end if
419 deallocate(this%w)
420 end if
421
422 if (allocated(this%r)) then
423 if (c_associated(this%r_d)) then
424 call device_unmap(this%r, this%r_d)
425 end if
426 deallocate(this%r)
427 end if
428
429 call this%schwarz%free()
430 call this%schwarz_mg%free()
431
432 call this%c_crs%free()
433 call this%c_mg%free()
434 call this%e%free()
435 call this%e_mg%free()
436 call this%e_crs%free()
437 call this%wf%free()
438
439 call this%gs_crs%free()
440 call this%gs_mg%free()
441 call this%interp_mid_crs%free()
442 call this%interp_fine_mid%free()
443
444 call this%bc_crs%free()
445 call this%bc_mg%free()
446 call this%bc_reg%free()
447
448 call this%bclst_reg%free()
449 call this%bclst_crs%free()
450 call this%bclst_mg%free()
451
452 if (allocated(this%crs_solver)) then
453 call this%crs_solver%free()
454 deallocate(this%crs_solver)
455 end if
456
457 if (allocated(this%amg_solver)) then
458 call this%amg_solver%free()
459 deallocate(this%amg_solver)
460 end if
461
462 if (allocated(this%pc_crs)) then
463 call precon_destroy(this%pc_crs)
464 end if
465
466 if (c_associated(this%hsmg_event)) then
467 call device_event_destroy(this%hsmg_event)
468 end if
469 if (c_associated(this%gs_event)) then
470 call device_event_destroy(this%gs_event)
471 end if
472
473 call this%dm_crs%free()
474 call this%dm_mg%free()
475 call this%Xh_crs%free()
476 call this%Xh_mg%free()
477
478 end subroutine hsmg_free
479
481 subroutine hsmg_solve(this, z, r, n)
482 integer, intent(in) :: n
483 class(hsmg_t), intent(inout) :: this
484 real(kind=rp), dimension(n), intent(inout) :: z
485 real(kind=rp), dimension(n), intent(inout) :: r
486 type(c_ptr) :: z_d, r_d
487 type(ksp_monitor_t) :: crs_info
488 integer :: thrdid, nthrds
489
490 call profiler_start_region('HSMG_solve', 8)
491 if (neko_bcknd_device .eq. 1) then
492 z_d = device_get_ptr(z)
493 r_d = device_get_ptr(r)
494 !We should not work with the input
495 call device_copy(this%r_d, r_d, n)
496 call this%bclst_reg%apply_scalar(this%r, n)
497
498 !OVERLAPPING Schwarz exchange and solve
499 !! DOWNWARD Leg of V-cycle, we are pretty hardcoded here but w/e
500 call device_col2(this%r_d, this%grids(3)%coef%mult_d, &
501 this%grids(3)%dof%size())
502 !Restrict to middle level
503 call this%interp_fine_mid%map(this%e%x, this%r, &
504 this%msh%nelv, this%grids(2)%Xh)
505 call this%grids(2)%gs_h%op(this%e%x, &
506 this%grids(2)%dof%size(), gs_op_add, this%gs_event)
507 call device_event_sync(this%gs_event)
508 !This should probably be double checked again
509 call device_copy(this%r_d, r_d, n)
510 call this%bclst_reg%apply_scalar(this%r, n)
511 call device_copy(this%w_d, this%e%x_d, this%grids(2)%dof%size())
512 call this%bclst_mg%apply_scalar(this%w, this%grids(2)%dof%size())
513 !OVERLAPPING Schwarz exchange and solve
514 call device_col2(this%w_d, this%grids(2)%coef%mult_d, &
515 this%grids(2)%dof%size())
516 !restrict residual to crs
517 call this%interp_mid_crs%map(this%wf%x, this%w, this%msh%nelv, &
518 this%grids(1)%Xh)
519 !Crs solve
520 call device_copy(this%w_d, this%e%x_d, this%grids(2)%dof%size())
521 call this%bclst_mg%apply_scalar(this%w, this%grids(2)%dof%size())
522
523 !$omp parallel private(thrdid, nthrds)
524
525 thrdid = 0
526 nthrds = 1
527 !$ thrdid = omp_get_thread_num()
528 !$ nthrds = omp_get_num_threads()
529
530 if (thrdid .eq. 0) then
531 call profiler_start_region('HSMG_schwarz', 9)
532 call this%grids(3)%schwarz%compute(z, this%r)
533 call this%grids(2)%schwarz%compute(this%grids(2)%e%x, this%w)
534 call profiler_end_region('HSMG_schwarz', 9)
535 end if
536 if (nthrds .eq. 1 .or. thrdid .eq. 1) then
537 call profiler_start_region('HSMG_coarse_grid', 10)
538 call this%grids(1)%gs_h%op(this%wf%x, &
539 this%grids(1)%dof%size(), gs_op_add, this%gs_event)
540 call device_event_sync(this%gs_event)
541 call this%grids(1)%bclst%apply_scalar(this%wf%x, &
542 this%grids(1)%dof%size())
543 call profiler_start_region('HSMG_coarse_solve', 11)
544 if (allocated(this%amg_solver)) then
545 call this%amg_solver%solve(this%grids(1)%e%x, this%wf%x, &
546 this%grids(1)%dof%size())
547 else
548 crs_info = this%crs_solver%solve(this%Ax, this%grids(1)%e, &
549 this%wf%x, &
550 this%grids(1)%dof%size(), &
551 this%grids(1)%coef, &
552 this%grids(1)%bclst, &
553 this%grids(1)%gs_h, this%niter)
554 end if
555 call profiler_end_region('HSMG_coarse_solve', 11)
556 call this%grids(1)%bclst%apply_scalar(this%grids(1)%e%x,&
557 this%grids(1)%dof%size())
558 call profiler_end_region('HSMG_coarse_grid', 10)
559 end if
560 !$omp end parallel
561
562 call this%interp_mid_crs%map(this%w, this%grids(1)%e%x, &
563 this%msh%nelv, this%grids(2)%Xh)
564 call device_add2(this%grids(2)%e%x_d, this%w_d, this%grids(2)%dof%size())
565
566 call this%interp_fine_mid%map(this%w, this%grids(2)%e%x, &
567 this%msh%nelv, this%grids(3)%Xh)
568 call device_add2(z_d, this%w_d, this%grids(3)%dof%size())
569 call this%grids(3)%gs_h%op(z, this%grids(3)%dof%size(), &
570 gs_op_add, this%gs_event)
571 call device_event_sync(this%gs_event)
572 call device_col2(z_d, this%grids(3)%coef%mult_d, &
573 this%grids(3)%dof%size())
574 else
575 !We should not work with the input
576 call copy(this%r, r, n)
577
578 !OVERLAPPING Schwarz exchange and solve
579 call this%grids(3)%schwarz%compute(z, this%r)
580 ! DOWNWARD Leg of V-cycle, we are pretty hardcoded here but w/e
581 call col2(this%r, this%grids(3)%coef%mult, &
582 this%grids(3)%dof%size())
583 !Restrict to middle level
584 call this%interp_fine_mid%map(this%w, this%r, &
585 this%msh%nelv, this%grids(2)%Xh)
586 call this%grids(2)%gs_h%op(this%w, this%grids(2)%dof%size(), gs_op_add)
587 !OVERLAPPING Schwarz exchange and solve
588 call this%grids(2)%schwarz%compute(this%grids(2)%e%x, this%w)
589 call col2(this%w, this%grids(2)%coef%mult, this%grids(2)%dof%size())
590 !restrict residual to crs
591 call this%interp_mid_crs%map(this%r, this%w, &
592 this%msh%nelv, this%grids(1)%Xh)
593 !Crs solve
594
595 call this%grids(1)%gs_h%op(this%r, this%grids(1)%dof%size(), gs_op_add)
596 call this%grids(1)%bclst%apply(this%r, this%grids(1)%dof%size())
597
598 call profiler_start_region('HSMG_coarse-solve', 11)
599 if (allocated(this%amg_solver)) then
600 call this%amg_solver%solve(this%grids(1)%e%x, this%r, &
601 this%grids(1)%dof%size())
602 else
603 crs_info = this%crs_solver%solve(this%Ax, this%grids(1)%e, this%r, &
604 this%grids(1)%dof%size(), &
605 this%grids(1)%coef, &
606 this%grids(1)%bclst, &
607 this%grids(1)%gs_h, this%niter)
608 end if
609 call profiler_end_region('HSMG_coarse-solve', 11)
610
611 call this%grids(1)%bclst%apply_scalar(this%grids(1)%e%x, &
612 this%grids(1)%dof%size())
613
614
615 call this%interp_mid_crs%map(this%w, this%grids(1)%e%x, &
616 this%msh%nelv, this%grids(2)%Xh)
617 call add2(this%grids(2)%e%x, this%w, this%grids(2)%dof%size())
618
619 call this%interp_fine_mid%map(this%w, this%grids(2)%e%x, &
620 this%msh%nelv, this%grids(3)%Xh)
621 call add2(z, this%w, this%grids(3)%dof%size())
622 call this%grids(3)%gs_h%op(z, this%grids(3)%dof%size(), gs_op_add)
623 call col2(z, this%grids(3)%coef%mult, this%grids(3)%dof%size())
624
625 end if
626 call profiler_end_region('HSMG_solve', 8)
627 end subroutine hsmg_solve
628end module hsmg
__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:113
Map a Fortran array to a device (allocate and associate)
Definition device.F90:83
Unmap a Fortran array from a device (deassociate and free)
Definition device.F90:89
Retrieves a parameter by name or assigns a provided default value. In the latter case also adds the m...
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
Coefficients.
Definition coef.f90:34
Jacobi preconditioner accelerator backend.
subroutine, public device_add2(a_d, b_d, n, strm)
Vector addition .
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 .
Device abstraction, common interface for various accelerators.
Definition device.F90:34
subroutine, public device_event_sync(event)
Synchronize an event.
Definition device.F90:1667
subroutine, public device_event_destroy(event)
Destroy a device event.
Definition device.F90:1623
subroutine, public device_event_create(event, flags)
Create a device event queue.
Definition device.F90:1589
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.
Krylov preconditioner.
Definition pc_hsmg.f90:61
subroutine hsmg_init(this, coef, bclst, hsmg_params)
Definition pc_hsmg.f90:148
subroutine hsmg_solve(this, z, r, n)
The h1mg preconditioner from Nek5000.
Definition pc_hsmg.f90:482
subroutine hsmg_set_h(this)
Definition pc_hsmg.f90:368
subroutine hsmg_free(this)
Definition pc_hsmg.f90:405
subroutine hsmg_fill_grid(dof, gs_h, xh, coef, bclst, schwarz, e, grids, l)
Definition pc_hsmg.f90:383
subroutine hsmg_init_from_components(this, coef, bclst, crs_solver, crs_pc, crs_monitor, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree)
Definition pc_hsmg.f90:193
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 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
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
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:50
Jacobi preconditioner SX-Aurora backend.
Implements multigrid using the TreeAMG hierarchy structure. USE:
Utilities.
Definition utils.f90:35
Defines a zero-valued Dirichlet boundary condition.
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
Coefficients defined on a given (mesh, ) tuple. Arrays use indices (i,j,k,e): element e,...
Definition coef.f90:63
Defines a jacobi preconditioner.
Generic Dirichlet boundary condition on .
Definition dirichlet.f90:49
Gather-scatter kernel.
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:64
Defines a jacobi preconditioner for SX-Aurora.
Type for the TreeAMG solver.
Zero-valued Dirichlet boundary condition. Used for no-slip walls, but also for various auxillary cond...
#define max(a, b)
Definition tensor.cu:40