56 use json_module,
only : json_file
68 use,
intrinsic :: iso_c_binding
75 integer :: smoother_itrs = 10
78 type(
gs_t),
pointer :: gs_h
99 class(
ax_t),
allocatable :: ax
105 integer :: last_metrics_version = -1
109 logical :: update_enabled = .false.
111 logical :: refresh_eigs = .true.
113 integer :: refresh_eigs_frequency = 20
115 logical :: eigs_warm_start = .true.
117 integer :: power_its_refresh = 20
119 integer :: n_refresh = 0
121 character(len=:),
allocatable :: cheby_acc
124 procedure, pass(this) :: init_from_components => &
135 class(
phmg_t),
intent(inout),
target :: this
136 type(
coef_t),
intent(in),
target :: coef
137 type(
bc_list_t),
intent(inout),
target :: bclst
138 type(json_file),
intent(inout) :: phmg_params
139 integer :: crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree
140 integer :: smoother_itrs
141 character(len=:),
allocatable :: cheby_acc
142 integer,
allocatable :: pcrs_sched(:)
143 logical :: update_enabled
158 crs_tamg_cheby_degree, 4)
160 if (phmg_params%valid_path(
'pcoarsening_schedule'))
then
161 call json_get(phmg_params,
'pcoarsening_schedule', pcrs_sched)
163 allocate(pcrs_sched(2))
169 update_enabled, .false.)
173 this%refresh_eigs, .true.)
175 this%refresh_eigs_frequency, 20)
177 this%eigs_warm_start, .true.)
179 this%power_its_refresh, 20)
181 call this%init_from_components(coef, bclst, smoother_itrs, &
182 cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree,&
183 pcrs_sched, update_enabled)
188 cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree, &
189 pcrs_sched, update_enabled)
190 class(
phmg_t),
intent(inout),
target :: this
191 type(
coef_t),
intent(in),
target :: coef
192 type(
bc_list_t),
intent(inout),
target :: bclst
193 integer,
intent(in) :: smoother_itrs
194 character(len=:),
allocatable :: cheby_acc
195 integer,
intent(in) :: crs_tamg_lvls, crs_tamg_itrs
196 integer,
intent(in) :: crs_tamg_cheby_degree
197 integer,
intent(in),
allocatable :: pcrs_sched(:)
198 logical,
intent(in),
optional :: update_enabled
199 integer :: lx_crs, lx_mid
200 integer,
allocatable :: lx_lvls(:)
201 integer :: n, i, j, st
202 class(
bc_t),
pointer :: bc_j
203 logical :: use_jacobi, use_cheby
208 this%cheby_acc = trim(cheby_acc)
210 if (
present(update_enabled))
then
211 this%update_enabled = update_enabled
213 this%update_enabled = .false.
216 this%nlvls =
size(pcrs_sched) + 1
217 allocate(lx_lvls(0:this%nlvls - 1))
218 lx_lvls(1:) = pcrs_sched + 1
220 allocate(this%phmg_hrchy%lvl(0:this%nlvls - 1))
222 this%phmg_hrchy%lvl(0)%lvl = 0
223 this%phmg_hrchy%lvl(0)%smoother_itrs = smoother_itrs
224 this%phmg_hrchy%lvl(0)%Xh => coef%Xh
225 this%phmg_hrchy%lvl(0)%coef => coef
226 this%phmg_hrchy%lvl(0)%dm_Xh => coef%dof
227 this%phmg_hrchy%lvl(0)%gs_h => coef%gs_h
229 do i = 1, this%nlvls - 1
230 allocate(this%phmg_hrchy%lvl(i)%Xh)
231 allocate(this%phmg_hrchy%lvl(i)%dm_Xh)
232 allocate(this%phmg_hrchy%lvl(i)%gs_h)
233 allocate(this%phmg_hrchy%lvl(i)%coef)
235 this%phmg_hrchy%lvl(i)%lvl = i
236 this%phmg_hrchy%lvl(i)%smoother_itrs = smoother_itrs
237 call this%phmg_hrchy%lvl(i)%Xh%init(
gll, lx_lvls(i), lx_lvls(i), &
239 call this%phmg_hrchy%lvl(i)%dm_Xh%init(coef%msh, &
240 this%phmg_hrchy%lvl(i)%Xh)
241 call this%phmg_hrchy%lvl(i)%gs_h%init(this%phmg_hrchy%lvl(i)%dm_Xh)
246 if (this%update_enabled)
then
247 call this%phmg_hrchy%lvl(i)%coef%init( &
250 call this%phmg_hrchy%lvl(i)%coef%init( &
255 do i = 0, this%nlvls - 1
256 call this%phmg_hrchy%lvl(i)%r%init(this%phmg_hrchy%lvl(i)%dm_Xh)
257 call this%phmg_hrchy%lvl(i)%w%init(this%phmg_hrchy%lvl(i)%dm_Xh)
258 call this%phmg_hrchy%lvl(i)%z%init(this%phmg_hrchy%lvl(i)%dm_Xh)
260 this%phmg_hrchy%lvl(i)%coef%ifh2 = coef%ifh2
261 call copy(this%phmg_hrchy%lvl(i)%coef%h1, coef%h1, &
262 this%phmg_hrchy%lvl(i)%dm_Xh%size())
264 call this%phmg_hrchy%lvl(i)%bc%init_base(this%phmg_hrchy%lvl(i)%coef)
265 if (bclst%size() .gt. 0 )
then
266 do j = 1, bclst%size()
268 call this%phmg_hrchy%lvl(i)%bc%mark_facets(bc_j%marked_facet)
271 call this%phmg_hrchy%lvl(i)%bc%finalize()
272 call this%phmg_hrchy%lvl(i)%bc%set_g(0.0_rp)
273 call this%phmg_hrchy%lvl(i)%bc_projector%mark(this%phmg_hrchy%lvl(i)%bc)
276 if (trim(cheby_acc) .eq.
"schwarz")
then
277 call this%phmg_hrchy%lvl(i)%schwarz%init( &
278 this%phmg_hrchy%lvl(i)%Xh, &
279 this%phmg_hrchy%lvl(i)%dm_Xh, &
280 this%phmg_hrchy%lvl(i)%gs_h, &
281 this%phmg_hrchy%lvl(i)%bc_projector, &
286 call this%phmg_hrchy%lvl(i)%device_jacobi%init(&
287 this%phmg_hrchy%lvl(i)%coef, &
288 this%phmg_hrchy%lvl(i)%dm_Xh, &
289 this%phmg_hrchy%lvl(i)%gs_h)
291 call this%phmg_hrchy%lvl(i)%jacobi%init(&
292 this%phmg_hrchy%lvl(i)%coef, &
293 this%phmg_hrchy%lvl(i)%dm_Xh, &
294 this%phmg_hrchy%lvl(i)%gs_h)
298 if (trim(cheby_acc) .eq.
"jacobi")
then
299 call this%phmg_hrchy%lvl(i)%cheby_device%init( &
300 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs, &
301 this%phmg_hrchy%lvl(i)%device_jacobi)
304 call this%phmg_hrchy%lvl(i)%cheby_device%init( &
305 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs)
307 if (trim(cheby_acc) .eq.
"schwarz")
then
308 this%phmg_hrchy%lvl(i)%cheby_device%schwarz => &
309 this%phmg_hrchy%lvl(i)%schwarz
314 if (trim(cheby_acc) .eq.
"jacobi")
then
315 call this%phmg_hrchy%lvl(i)%cheby%init( &
316 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs, &
317 this%phmg_hrchy%lvl(i)%jacobi)
320 call this%phmg_hrchy%lvl(i)%cheby%init( &
321 this%phmg_hrchy%lvl(i)%dm_Xh%size(), smoother_itrs)
323 if (trim(cheby_acc) .eq.
"schwarz")
then
324 this%phmg_hrchy%lvl(i)%cheby%schwarz => &
325 this%phmg_hrchy%lvl(i)%schwarz
336 call ax_helm_allocator(this%ax, type_name =
"standard")
339 allocate(this%intrp(this%nlvls - 1))
340 do i = 1, this%nlvls -1
341 call this%intrp(i)%init(this%phmg_hrchy%lvl(i-1)%Xh, &
342 this%phmg_hrchy%lvl(i)%Xh)
346 if (this%update_enabled)
then
347 allocate(this%crd_intrp(this%nlvls - 1))
348 do i = 1, this%nlvls - 1
349 call this%crd_intrp(i)%init(this%phmg_hrchy%lvl(i)%Xh, &
350 this%phmg_hrchy%lvl(0)%Xh)
354 call this%amg_solver%init(this%ax, this%phmg_hrchy%lvl(this%nlvls -1)%Xh, &
355 this%phmg_hrchy%lvl(this%nlvls -1)%coef, this%msh, &
356 this%phmg_hrchy%lvl(this%nlvls-1)%gs_h, crs_tamg_lvls, &
357 this%phmg_hrchy%lvl(this%nlvls -1)%bc_projector, &
358 crs_tamg_itrs, crs_tamg_cheby_degree)
361 this%last_metrics_version = this%phmg_hrchy%lvl(0)%coef%metrics_version
364 if (this%update_enabled)
then
365 do i = 0, this%nlvls - 1
366 this%phmg_hrchy%lvl(i)%cheby%warm_start_eigs = this%eigs_warm_start
367 this%phmg_hrchy%lvl(i)%cheby%power_its_refresh = &
368 this%power_its_refresh
369 this%phmg_hrchy%lvl(i)%cheby_device%warm_start_eigs = &
371 this%phmg_hrchy%lvl(i)%cheby_device%power_its_refresh = &
372 this%power_its_refresh
374 call this%amg_solver%set_eig_refresh(this%eigs_warm_start, &
375 this%power_its_refresh)
377 if (trim(cheby_acc) .eq.
"schwarz")
then
378 call neko_warning(
"PHMG: the Schwarz smoother is not refreshed " // &
379 "when the mesh changes. Its local solves stay at the " // &
387 class(
phmg_t),
intent(inout) :: this
390 call this%amg_solver%free()
392 if (
allocated(this%intrp))
then
393 do i = 1,
size(this%intrp)
394 call this%intrp(i)%free()
396 deallocate(this%intrp)
399 if (
allocated(this%crd_intrp))
then
400 do i = 1,
size(this%crd_intrp)
401 call this%crd_intrp(i)%free()
403 deallocate(this%crd_intrp)
406 if (
allocated(this%ax))
then
410 if (
allocated(this%phmg_hrchy%lvl))
then
411 do i = lbound(this%phmg_hrchy%lvl, 1), ubound(this%phmg_hrchy%lvl, 1)
412 call this%phmg_hrchy%lvl(i)%r%free()
413 call this%phmg_hrchy%lvl(i)%w%free()
414 call this%phmg_hrchy%lvl(i)%z%free()
416 call this%phmg_hrchy%lvl(i)%cheby%free()
417 call this%phmg_hrchy%lvl(i)%cheby_device%free()
418 call this%phmg_hrchy%lvl(i)%jacobi%free()
419 call this%phmg_hrchy%lvl(i)%device_jacobi%free()
421 if (
allocated(this%phmg_hrchy%lvl(i)%schwarz%work1))
then
422 call this%phmg_hrchy%lvl(i)%schwarz%free()
425 call this%phmg_hrchy%lvl(i)%bc_projector%free()
426 call this%phmg_hrchy%lvl(i)%bc%free()
430 if (this%phmg_hrchy%lvl(i)%lvl .gt. 0)
then
431 call this%phmg_hrchy%lvl(i)%coef%free()
432 call this%phmg_hrchy%lvl(i)%gs_h%free()
433 call this%phmg_hrchy%lvl(i)%dm_Xh%free()
434 call this%phmg_hrchy%lvl(i)%Xh%free()
435 deallocate(this%phmg_hrchy%lvl(i)%coef)
436 deallocate(this%phmg_hrchy%lvl(i)%gs_h)
437 deallocate(this%phmg_hrchy%lvl(i)%dm_Xh)
438 deallocate(this%phmg_hrchy%lvl(i)%Xh)
441 nullify(this%phmg_hrchy%lvl(i)%coef)
442 nullify(this%phmg_hrchy%lvl(i)%gs_h)
443 nullify(this%phmg_hrchy%lvl(i)%dm_Xh)
444 nullify(this%phmg_hrchy%lvl(i)%Xh)
446 deallocate(this%phmg_hrchy%lvl)
454 class(
phmg_t),
intent(inout) :: this
455 integer,
intent(in) :: n
456 real(kind=
rp),
dimension(n),
intent(inout) :: z
457 real(kind=
rp),
dimension(n),
intent(inout) :: r
458 type(c_ptr) :: z_d, r_d
463 associate( mglvl => this%phmg_hrchy%lvl)
484 mglvl(0)%r%x(i,1,1,1) = r(i)
486 mglvl(0)%z%x(i,1,1,1) = 0.0_rp
487 mglvl(0)%w%x(i,1,1,1) = 0.0_rp
499 z(i) = mglvl(0)%z%x(i,1,1,1)
510 class(
phmg_t),
intent(inout) :: this
511 integer :: fine_version
514 if (.not. this%update_enabled)
return
516 fine_version = this%phmg_hrchy%lvl(0)%coef%metrics_version
519 if (fine_version .eq. this%last_metrics_version)
return
525 do_eigs = (this%refresh_eigs) .and. &
526 (this%refresh_eigs_frequency .gt. 0) .and. &
527 (mod(this%n_refresh,
max(this%refresh_eigs_frequency, 1)) &
531 this%n_refresh = this%n_refresh + 1
532 this%last_metrics_version = fine_version
540 class(
phmg_t),
intent(inout) :: this
543 if (.not.
allocated(this%crd_intrp))
then
544 call neko_error(
"PHMG: update requested but coordinate " // &
545 "interpolators were not initialized.")
548 call profiler_start_region(
'PHMG_update_geometry')
549 associate(mg => this%phmg_hrchy%lvl, nelv => this%msh%nelv)
550 do i = 1, this%nlvls - 1
551 call this%crd_intrp(i)%map(mg(i)%dm_Xh%x, mg(0)%dm_Xh%x, &
553 call this%crd_intrp(i)%map(mg(i)%dm_Xh%y, mg(0)%dm_Xh%y, &
555 call this%crd_intrp(i)%map(mg(i)%dm_Xh%z, mg(0)%dm_Xh%z, &
558 call mg(i)%coef%recompute_metrics()
561 call profiler_end_region(
'PHMG_update_geometry')
569 class(
phmg_t),
intent(inout) :: this
572 call profiler_start_region(
'PHMG_update_smoother_acc')
573 associate(mg => this%phmg_hrchy%lvl)
574 select case (trim(this%cheby_acc))
576 do i = 0, this%nlvls - 1
577 if (neko_bcknd_device .eq. 1)
then
578 call mg(i)%device_jacobi%update()
580 call mg(i)%jacobi%update()
590 call profiler_end_region(
'PHMG_update_smoother_acc')
597 class(
phmg_t),
intent(inout) :: this
600 associate(mg => this%phmg_hrchy%lvl)
601 do i = 0, this%nlvls - 1
602 if (neko_bcknd_device .eq. 1)
then
603 mg(i)%cheby_device%recompute_eigs = .true.
605 mg(i)%cheby%recompute_eigs = .true.
610 call this%amg_solver%invalidate_eigs()
616 class(
phmg_t),
intent(inout) :: this
617 type(ksp_monitor_t) :: ksp_results
618 character(len=2) :: lvl_name
621 associate(mg => this%phmg_hrchy%lvl, intrp => this%intrp, &
622 msh => this%msh, ax => this%Ax)
623 do lvl = 0, this%nlvls-2
624 write(lvl_name,
'(I0)') lvl
625 call profiler_start_region(
"PHMG_level_" // trim(lvl_name))
626 associate(z => mg(lvl)%z, r => mg(lvl)%r, w => mg(lvl)%w)
630 if (neko_bcknd_device .eq. 1)
then
631 mg(lvl)%cheby_device%zero_initial_guess = .true.
632 ksp_results = mg(lvl)%cheby_device%solve(ax, z, &
633 r%x, mg(lvl)%dm_Xh%size(), &
634 mg(lvl)%coef, mg(lvl)%bc_projector, &
635 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
637 mg(lvl)%cheby%zero_initial_guess = .true.
638 ksp_results = mg(lvl)%cheby%solve(ax, z, &
639 r%x, mg(lvl)%dm_Xh%size(), &
640 mg(lvl)%coef, mg(lvl)%bc_projector, &
641 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
647 call ax%compute(w%x, z%x, mg(lvl)%coef, msh, mg(lvl)%Xh)
648 call mg(lvl)%gs_h%op(w%x, mg(lvl)%dm_Xh%size(), gs_op_add, &
650 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
651 call mg(lvl)%bc_projector%apply(w%x, mg(lvl)%dm_Xh%size())
653 if (neko_bcknd_device .eq. 1)
then
654 call device_add2s1(w%x_d, r%x_d, -1.0_rp, mg(lvl)%dm_Xh%size())
661 do i = 1, mg(lvl)%dm_Xh%size()
662 w%x(i,1,1,1) = r%x(i,1,1,1) - w%x(i,1,1,1)
670 if (neko_bcknd_device .eq. 1)
then
671 call device_col2(w%x_d, mg(lvl)%coef%mult_d, mg(lvl)%dm_Xh%size())
673 call col2(w%x, mg(lvl)%coef%mult, mg(lvl)%dm_Xh%size())
676 call intrp(lvl+1)%map(mg(lvl+1)%r%x, w%x, msh%nelv, mg(lvl+1)%Xh)
678 call mg(lvl+1)%gs_h%op(mg(lvl+1)%r%x, mg(lvl+1)%dm_Xh%size(), &
679 gs_op_add, glb_cmd_event)
680 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
682 call mg(lvl+1)%bc_projector%apply( &
684 mg(lvl+1)%dm_Xh%size())
686 if (neko_bcknd_device .eq. 1)
then
687 call device_rzero(mg(lvl+1)%z%x_d, mg(lvl+1)%dm_Xh%size())
694 do i = 1, mg(lvl+1)%dm_Xh%size()
695 mg(lvl+1)%z%x(i,1,1,1) = 0.0_rp
700 call profiler_end_region(
"PHMG_level_" // trim(lvl_name))
703 call profiler_start_region(
'PHMG_coarse-solve' )
707 call this%amg_solver%solve(mg(this%nlvls-1)%z%x, &
708 mg(this%nlvls-1)%r%x, &
709 mg(this%nlvls-1)%dm_Xh%size())
710 call profiler_end_region(
'PHMG_coarse-solve' )
712 do lvl = (this%nlvls-2), 0, -1
713 write(lvl_name,
'(I0)') lvl
714 call profiler_start_region(
"PHMG_level_" // trim(lvl_name))
715 associate(z => mg(lvl)%z, r => mg(lvl)%r, w => mg(lvl)%w)
719 call intrp(lvl+1)%map(w%x, mg(lvl+1)%z%x, msh%nelv, mg(lvl)%Xh)
721 call mg(lvl)%gs_h%op(w%x, mg(lvl)%dm_Xh%size(), gs_op_add, &
723 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
725 if (neko_bcknd_device .eq. 1)
then
726 call device_col2(w%x_d, mg(lvl)%coef%mult_d, mg(lvl)%dm_Xh%size())
728 call col2(w%x, mg(lvl)%coef%mult, mg(lvl)%dm_Xh%size())
734 if (neko_bcknd_device .eq. 1)
then
735 call device_add2(z%x_d, w%x_d, mg(lvl)%dm_Xh%size())
742 do i = 1, mg(lvl)%dm_Xh%size()
743 z%x(i,1,1,1) = z%x(i,1,1,1) + w%x(i,1,1,1)
751 if (neko_bcknd_device .eq. 1)
then
752 ksp_results = mg(lvl)%cheby_device%solve(ax, z, &
753 r%x, mg(lvl)%dm_Xh%size(), &
754 mg(lvl)%coef, mg(lvl)%bc_projector, &
755 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
757 ksp_results = mg(lvl)%cheby%solve(ax, z, &
758 r%x, mg(lvl)%dm_Xh%size(), &
759 mg(lvl)%coef, mg(lvl)%bc_projector, &
760 mg(lvl)%gs_h, niter = mg(lvl)%smoother_itrs)
763 call profiler_end_region(
"PHMG_level_" // trim(lvl_name))
780 class(ax_t),
intent(inout) :: Ax
781 type(mesh_t),
intent(inout) :: msh
782 type(field_t),
intent(inout) :: z, r, w
783 integer,
intent(in) :: n, lvl
784 integer :: i, j, iblk, ni, niblk
786 ni = mg%smoother_itrs
787 if (neko_bcknd_device .eq. 1)
then
789 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
790 call mg%gs_h%op(w%x, n, gs_op_add, glb_cmd_event)
791 call device_stream_wait_event(glb_cmd_queue, glb_cmd_event, 0)
792 call mg%bc_projector%apply(w%x, n)
793 call device_add2s1(w%x_d, r%x_d, -1.0_rp, n)
795 call mg%device_jacobi%solve(w%x, w%x, n)
797 call device_add2s2(z%x_d, w%x_d, 0.6_rp, n)
801 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
802 call mg%gs_h%op(w%x, n, gs_op_add)
803 call mg%bc_projector%apply(w%x, n)
804 call add2s1(w%x, r%x, -1.0_rp, n)
806 call mg%jacobi%solve(w%x, w%x, n)
808 call add2s2(z%x, w%x, 0.6_rp, n)
817 class(ax_t),
intent(inout) :: Ax
818 type(mesh_t),
intent(inout) :: msh
819 type(field_t) :: z, r, w
821 character(len=LOG_SIZE) :: log_buf
822 call ax%compute(w%x, z%x, mg%coef, msh, mg%Xh)
823 call mg%gs_h%op(w%x, mg%dm_Xh%size(), gs_op_add)
824 call mg%bc_projector%apply(w%x, mg%dm_Xh%size())
825 call device_add2s1(w%x_d, r%x_d, -1.0_rp, mg%dm_Xh%size())
826 val = device_glsc2(w%x_d, w%x_d, mg%dm_Xh%size())
828 write(log_buf,
'(A15,I4,F12.6)')
'PRESMOO - PRE', lvl, val
829 else if (typ .eq. 2)
then
830 write(log_buf,
'(A15,I4,F12.6)')
'PRESMOO -POST', lvl, val
831 else if (typ .eq. 3)
then
832 write(log_buf,
'(A15,I4,F12.6)')
'POSTSMOO- PRE', lvl, val
833 else if (typ .eq. 4)
then
834 write(log_buf,
'(A15,I4,F12.6)')
'POSTSMOO-POST', lvl, val
835 else if (typ .eq. 5)
then
836 write(log_buf,
'(A15,I4,F12.6)')
'TAMG - PRE', lvl, val
837 else if (typ .eq. 6)
then
838 write(log_buf,
'(A15,I4,F12.6)')
'TAMG -POST', lvl, val
840 write(log_buf,
'(A15,I4,F12.6)')
'RESID', lvl, val
842 call neko_log%message(log_buf)
846 integer,
intent(in) :: nlvls
847 integer,
intent(in) :: smoo_type
850 character(len=LOG_SIZE) :: log_buf, smoo_name
852 call neko_log%section(
'PHMG')
854 if (smoo_type .eq. 1)
then
855 write(smoo_name,
'(A16)')
'CHEBY-acc JACOBI'
856 else if (smoo_type .eq. 2)
then
857 write(smoo_name,
'(A17)')
'CHEBY-acc SCHWARZ'
859 write(smoo_name,
'(A5)')
'CHEBY'
862 write(log_buf,
'(A28,I2,A8)') &
863 'Creating PHMG hierarchy with', &
865 call neko_log%message(log_buf)
869 write(log_buf,
'(A8,I2,A8,I2)') &
870 '-- level', i,
'-- lx:',
phmg%lvl(i)%Xh%lx
871 call neko_log%message(log_buf)
873 if (i .eq. clvl)
then
874 write(log_buf,
'(A19,A20)') &
876 call neko_log%message(log_buf)
878 write(log_buf,
'(A22,A20)') &
881 call neko_log%message(log_buf)
883 write(log_buf,
'(A28,I2)') &
885 phmg%lvl(i)%smoother_itrs
886 call neko_log%message(log_buf)
890 call neko_log%end_section()
__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.
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.
Defines a boundary condition.
Chebyshev preconditioner.
Chebyshev preconditioner.
integer, parameter, public coef_full
Retain every coefficient. The default, and the only scope that supports recompute_metrics(),...
integer, parameter, public coef_operator
Retain only what applying a discrete operator needs: , h1, h2, B and mult.
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.
subroutine, public device_stream_wait_event(stream, event, flags)
Synchronize a device stream with an event.
type(c_ptr), bind(C), public glb_cmd_queue
Global command queue.
type(c_ptr), bind(C), public glb_cmd_event
Event for the global command queue.
Defines a dirichlet boundary condition.
Defines a mapping of the degrees of freedom.
Routines to interpolate between different spaces.
Utilities for retrieving parameters from the case files.
Implements the base abstract type for Krylov solvers plus helper types.
integer, parameter, public ksp_max_iter
Maximum number of iters.
type(log_t), public neko_log
Global log stream.
integer, parameter, public log_size
subroutine, public add2s1(a, b, c1, n)
Vector addition with scalar multiplication (multiplication on first argument)
subroutine, public add2(a, b, n)
Vector addition .
subroutine, public col2(a, b, n)
Vector multiplication .
subroutine, public copy(a, b, n)
Copy a vector .
subroutine, public add2s2(a, b, c1, n)
Vector addition with scalar multiplication (multiplication on second argument)
integer, parameter neko_bcknd_device
integer, parameter, public rp
Global precision used in computations.
Hybrid ph-multigrid preconditioner.
subroutine phmg_update_smoother_acc(this)
Rebuild the accelerator the Chebyshev smoother uses, which depends on the geometry.
subroutine phmg_jacobi_smoother(z, r, w, mg, msh, ax, n, lvl)
Wraps jacobi solve as a residual update relaxation method.
subroutine phmg_solve(this, z, r, n)
subroutine phmg_init(this, coef, bclst, phmg_params)
subroutine phmg_update(this)
Bring the preconditioner back in sync after the mesh has changed.
subroutine phmg_update_smoother_eigs(this)
Mark every smoother to re-estimate its eigenvalues on the next solve.
subroutine phmg_update_coarse_geometry(this)
Sample the new fine coordinates on the coarse levels and rebuild their metrics. Level 0 shares the ca...
subroutine phmg_free(this)
subroutine phmg_mg_cycle(this)
subroutine phmg_resid_monitor(z, r, w, mg, msh, ax, lvl, typ)
subroutine print_phmg_info(nlvls, smoo_type, phmg)
subroutine phmg_init_from_components(this, coef, bclst, smoother_itrs, cheby_acc, crs_tamg_lvls, crs_tamg_itrs, crs_tamg_cheby_degree, pcrs_sched, update_enabled)
subroutine, public profiler_start_region(name, region_id)
Started a named (name) profiler region.
subroutine, public profiler_end_region(name, region_id)
End the most recently started profiler region.
Implements scalar_projector_t.
Overlapping schwarz solves.
Defines a function space.
integer, parameter, public gll
Implements multigrid using the TreeAMG hierarchy structure. USE:
subroutine, public neko_warning(warning_msg)
Reports a warning to standard output.
Base type for a matrix-vector product providing .
Base type for a boundary condition.
A list of allocatable `bc_t`. Follows the standard interface of lists.
Defines a Chebyshev preconditioner.
Defines a Chebyshev preconditioner.
Coefficients defined on a given (mesh, ) tuple. Arrays use indices (i,j,k,e): element e,...
Defines a jacobi preconditioner.
Generic Dirichlet boundary condition on .
Interpolation between two space::space_t.
Defines a jacobi preconditioner.
Type for storing initial and final residuals in a Krylov solver.
Base abstract type for a canonical Krylov method, solving .
Defines a canonical Krylov preconditioner.
Projector for scalar boundary conditions.
The function space for the SEM solution fields.
Type for the TreeAMG solver.