Neko 1.99.9
A portable framework for high-order spectral element flow simulations
Loading...
Searching...
No Matches
tree_amg_smoother.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!
35 use tree_amg, only : tamg_hierarchy_t
37 use num_types, only : rp
38 use math, only : col2, add2, add2s2, glsc2, glsc3, sub2, cmult, &
43 use krylov, only : ksp_monitor_t
45 use gather_scatter, only : gs_t, gs_op_add
46 use logger, only : neko_log, log_size
52 use, intrinsic :: iso_c_binding
53 implicit none
54 private
55
57 type, public :: amg_jacobi_t
58 real(kind=rp), allocatable :: d(:)
59 real(kind=rp), allocatable :: w(:)
60 real(kind=rp), allocatable :: r(:)
61 real(kind=rp) :: omega
62 integer :: lvl
63 integer :: n
64 integer :: max_iter = 10
65 logical :: recompute_diag = .true.
66 contains
67 procedure, pass(this) :: init => amg_jacobi_init
68 procedure, pass(this) :: solve => amg_jacobi_solve
69 procedure, pass(this) :: comp_diag => amg_jacobi_diag
70 procedure, pass(this) :: free => amg_jacobi_free
71 end type amg_jacobi_t
72
74 type, public :: amg_cheby_t
75 real(kind=rp), allocatable :: d(:)
76 type(c_ptr) :: d_d = c_null_ptr
77 real(kind=rp), allocatable :: w(:)
78 type(c_ptr) :: w_d = c_null_ptr
79 real(kind=rp), allocatable :: r(:)
80 type(c_ptr) :: r_d = c_null_ptr
83 real(kind=rp), allocatable :: ev(:)
84 type(c_ptr) :: ev_d = c_null_ptr
85 real(kind=rp) :: tha, dlt
86 integer :: lvl
87 integer :: n
88 integer :: power_its = 250
90 integer :: power_its_refresh = 20
92 logical :: warm_start_eigs = .false.
94 logical :: eigs_computed = .false.
95 integer :: max_iter = 10
96 logical :: recompute_eigs = .true.
97 contains
98 procedure, pass(this) :: init => amg_cheby_init
99 procedure, pass(this) :: solve => amg_cheby_solve
100 procedure, pass(this) :: comp_eig => amg_cheby_power
101 procedure, pass(this) :: device_solve => amg_device_cheby_solve
102 procedure, pass(this) :: device_comp_eig => amg_device_cheby_power
103 procedure, pass(this) :: free => amg_cheby_free
104 end type amg_cheby_t
105
106contains
107
113 subroutine amg_cheby_init(this, n, lvl, max_iter)
114 class(amg_cheby_t), intent(inout), target :: this
115 integer, intent(in) :: n
116 integer, intent(in) :: lvl
117 integer, intent(in) :: max_iter
118
119 allocate(this%d(n))
120 allocate(this%w(n))
121 allocate(this%r(n))
122 if (neko_bcknd_device .eq. 1) then
123 call device_map(this%d, this%d_d, n)
124 call device_map(this%w, this%w_d, n)
125 call device_map(this%r, this%r_d, n)
126 end if
127 this%n = n
128 this%lvl = lvl
129 this%max_iter = max_iter
130 this%recompute_eigs = .true.
131
132 call amg_smoo_monitor(lvl, this)
133
134 end subroutine amg_cheby_init
135
137 subroutine amg_cheby_free(this)
138 class(amg_cheby_t), intent(inout), target :: this
139 if (allocated(this%d)) then
140 if (neko_bcknd_device .eq. 1 .and. c_associated(this%d_d)) then
141 call device_unmap(this%d, this%d_d)
142 end if
143 deallocate(this%d)
144 end if
145 if (allocated(this%w)) then
146 if (neko_bcknd_device .eq. 1 .and. c_associated(this%w_d)) then
147 call device_unmap(this%w, this%w_d)
148 end if
149 deallocate(this%w)
150 end if
151 if (allocated(this%r)) then
152 if (neko_bcknd_device .eq. 1 .and. c_associated(this%r_d)) then
153 call device_unmap(this%r, this%r_d)
154 end if
155 deallocate(this%r)
156 end if
157 if (allocated(this%ev)) then
158 if (neko_bcknd_device .eq. 1 .and. c_associated(this%ev_d)) then
159 call device_unmap(this%ev, this%ev_d)
160 end if
161 deallocate(this%ev)
162 end if
163 end subroutine amg_cheby_free
164
165
169 subroutine amg_cheby_power(this, amg, n)
170 class(amg_cheby_t), intent(inout) :: this
171 type(tamg_hierarchy_t), intent(inout) :: amg
172 integer, intent(in) :: n
173 real(kind=rp) :: lam, b, a, rn
174 real(kind=rp), parameter :: boost = 1.1_rp
175 real(kind=rp), parameter :: lam_factor = 30.0_rp
176 real(kind=rp) :: wtw, dtw, dtd
177 integer, allocatable :: fixed_seed(:), saved_seed(:)
178 integer :: i, rnd_n, its
179 logical :: warm
180
181 call profiler_start_region('AMG_cheby_power')
182 warm = this%warm_start_eigs .and. this%eigs_computed .and. &
183 allocated(this%ev)
184 associate(w => this%w, d => this%d, coef => amg%coef, gs_h => amg%gs_h, &
185 msh => amg%msh, xh => amg%Xh, bc_projector => amg%bc_projector)
186
187 if (warm) then
188 its = this%power_its_refresh
189 call copy(d, this%ev, n)
190 else
191 its = this%power_its
192
193 ! Save current random seed and set a fixed seed
194 call random_seed(size = rnd_n)
195 allocate(saved_seed(rnd_n))
196 allocate(fixed_seed(rnd_n))
197 fixed_seed = 3901
198 call random_seed(get = saved_seed)
199 call random_seed(put = fixed_seed)
200
201 do i = 1, n
202 call random_number(rn)
203 d(i) = rn + 10.0_rp
204 end do
205
206 ! Restore saved random seed
207 call random_seed(put = saved_seed)
208
209 if (this%lvl .eq. 0) then
210 call gs_h%op(d, n, gs_op_add)!TODO
211 call bc_projector%apply(d, n)
212 end if
213 end if
214 !Power method to get lamba max
215 do i = 1, its
216 call amg%matvec(w, d, this%lvl)
217
218 if (this%lvl .eq. 0) then
219 wtw = glsc3(w, coef%mult, w, n)
220 else
221 wtw = glsc2(w, w, n)
222 end if
223
224 call cmult2(d, w, 1.0_rp/sqrt(wtw), n)
225 end do
226
227 call amg%matvec(w, d, this%lvl)
228
229 if (this%lvl .eq. 0) then
230 dtw = glsc3(d, coef%mult, w, n)
231 dtd = glsc3(d, coef%mult, d, n)
232 else
233 dtw = glsc2(d, w, n)
234 dtd = glsc2(d, d, n)
235 end if
236 lam = dtw / dtd
237 b = lam * boost
238 a = lam / lam_factor
239 this%tha = (b+a)/2.0_rp
240 this%dlt = (b-a)/2.0_rp
241
242 if (this%warm_start_eigs) then
243 if (.not. allocated(this%ev)) then
244 allocate(this%ev(this%n))
245 end if
246 call copy(this%ev, d, n)
247 end if
248 this%eigs_computed = .true.
249
250 this%recompute_eigs = .false.
251 call amg_cheby_monitor(this%lvl, lam)
252 end associate
253 call profiler_end_region('AMG_cheby_power')
254 end subroutine amg_cheby_power
255
262 subroutine amg_cheby_solve(this, x, f, n, amg, zero_init)
263 class(amg_cheby_t), intent(inout) :: this
264 integer, intent(in) :: n
265 real(kind=rp), dimension(n), intent(inout) :: x
266 real(kind=rp), dimension(n), intent(inout) :: f
267 class(tamg_hierarchy_t), intent(inout) :: amg
268 type(ksp_monitor_t) :: ksp_results
269 logical, optional, intent(in) :: zero_init
270 integer :: iter, max_iter, i
271 real(kind=rp) :: rtr, rnorm
272 real(kind=rp) :: rhok, rhokp1, s1, thet, delt, tmp1, tmp2
273 logical :: zero_initial_guess
274
275 if (this%recompute_eigs) then
276 call this%comp_eig(amg, n)
277 end if
278 if (present(zero_init)) then
279 zero_initial_guess = zero_init
280 else
281 zero_initial_guess = .false.
282 end if
283 max_iter = this%max_iter
284
285 associate(w => this%w, r => this%r, d => this%d, &
286 bc_projector => amg%bc_projector)
287 call copy(r, f, n)
288 if (.not. zero_initial_guess) then
289 call amg%matvec(w, x, this%lvl)
290 call sub2(r, w, n)
291 end if
292
293 thet = this%tha
294 delt = this%dlt
295 s1 = thet / delt
296 rhok = 1.0_rp / s1
297
298 ! First iteration
299 !OCL NORECURRENCE, NOVREC, NOALIAS
300 !DIR$ CONCURRENT
301 !DIR$ IVDEP
302 !GCC$ ivdep
303 !$omp parallel do
304 do i = 1, n
305 d(i) = 1.0_rp/thet * r(i)
306 x(i) = x(i) + d(i)
307 end do
308 !$omp end parallel do
309
310 ! Rest of iterations
311 do iter = 2, max_iter
312 call amg%matvec(w, d, this%lvl)
313
314 rhokp1 = 1.0_rp / (2.0_rp * s1 - rhok)
315 tmp1 = rhokp1 * rhok
316 tmp2 = 2.0_rp * rhokp1 / delt
317 rhok = rhokp1
318
319 !$omp parallel private(i)
320 !OCL NORECURRENCE, NOVREC, NOALIAS
321 !DIR$ CONCURRENT
322 !DIR$ IVDEP
323 !GCC$ ivdep
324 !$omp do
325 do i = 1, n
326 r(i) = r(i) - w(i)
327 d(i) = tmp1 * d(i) + tmp2 * r(i)
328 x(i) = x(i) + d(i)
329 end do
330 !$omp end do
331 !$omp end parallel
332 end do
333 end associate
334 end subroutine amg_cheby_solve
335
339 subroutine amg_device_cheby_power(this, amg, n)
340 class(amg_cheby_t), intent(inout) :: this
341 type(tamg_hierarchy_t), intent(inout) :: amg
342 integer, intent(in) :: n
343 real(kind=rp) :: lam, b, a, rn
344 real(kind=rp), parameter :: boost = 1.1_rp
345 real(kind=rp), parameter :: lam_factor = 30.0_rp
346 real(kind=rp) :: wtw, dtw, dtd
347 integer, allocatable :: fixed_seed(:), saved_seed(:)
348 integer :: i, rnd_n, its
349 logical :: warm
350
351 call profiler_start_region('AMG_cheby_power')
352 warm = this%warm_start_eigs .and. this%eigs_computed .and. &
353 c_associated(this%ev_d)
354 associate(w => this%w, d => this%d, coef => amg%coef, gs_h => amg%gs_h, &
355 msh => amg%msh, xh => amg%Xh, bc_projector => amg%bc_projector)
356
357 if (warm) then
358 its = this%power_its_refresh
359 call device_copy(this%d_d, this%ev_d, n)
360 else
361 its = this%power_its
362
363 ! Save current random seed and set a fixed seed
364 call random_seed(size = rnd_n)
365 allocate(saved_seed(rnd_n))
366 allocate(fixed_seed(rnd_n))
367 fixed_seed = 3901
368 call random_seed(get = saved_seed)
369 call random_seed(put = fixed_seed)
370
371 do i = 1, n
372 call random_number(rn)
373 d(i) = rn + 10.0_rp
374 end do
375 call device_memcpy(this%d, this%d_d, n, host_to_device, .true.)
376
377 ! Restore saved random seed
378 call random_seed(put = saved_seed)
379
380 if (this%lvl .eq. 0) then
381 call gs_h%op(d, n, gs_op_add)!TODO
382 call bc_projector%apply(d, n)
383 end if
384 end if
385 do i = 1, its
386 call amg%device_matvec(w, d, this%w_d, this%d_d, this%lvl)
387
388 if (this%lvl .eq. 0) then
389 wtw = device_glsc3(this%w_d, coef%mult_d, this%w_d, n)
390 else
391 wtw = device_glsc2(this%w_d, this%w_d, n)
392 end if
393
394 call device_cmult2(this%d_d, this%w_d, 1.0_rp/sqrt(wtw), n)
395 end do
396
397 call amg%device_matvec(w, d, this%w_d, this%d_d, this%lvl)
398
399 if (this%lvl .eq. 0) then
400 dtw = device_glsc3(this%d_d, coef%mult_d, this%w_d, n)
401 dtd = device_glsc3(this%d_d, coef%mult_d, this%d_d, n)
402 else
403 dtw = device_glsc2(this%d_d, this%w_d, n)
404 dtd = device_glsc2(this%d_d, this%d_d, n)
405 end if
406 lam = dtw / dtd
407 b = lam * boost
408 a = lam / lam_factor
409 this%tha = (b+a)/2.0_rp
410 this%dlt = (b-a)/2.0_rp
411
412 if (this%warm_start_eigs) then
413 if (.not. c_associated(this%ev_d)) then
414 allocate(this%ev(this%n))
415 call device_map(this%ev, this%ev_d, this%n)
416 end if
417 call device_copy(this%ev_d, this%d_d, n)
418 end if
419 this%eigs_computed = .true.
420
421 this%recompute_eigs = .false.
422 call amg_cheby_monitor(this%lvl, lam)
423 end associate
424 call profiler_end_region('AMG_cheby_power')
425 end subroutine amg_device_cheby_power
426
433 subroutine amg_device_cheby_solve(this, x, f, x_d, f_d, n, amg, zero_init)
434 class(amg_cheby_t), intent(inout) :: this
435 integer, intent(in) :: n
436 real(kind=rp), dimension(n), intent(inout) :: x
437 real(kind=rp), dimension(n), intent(inout) :: f
438 type(c_ptr) :: x_d
439 type(c_ptr) :: f_d
440 class(tamg_hierarchy_t), intent(inout) :: amg
441 type(ksp_monitor_t) :: ksp_results
442 logical, optional, intent(in) :: zero_init
443 integer :: iter, max_iter
444 real(kind=rp) :: rtr, rnorm
445 real(kind=rp) :: rhok, rhokp1, s1, thet, delt, tmp1, tmp2
446 logical :: zero_initial_guess
447
448 if (this%recompute_eigs) then
449 call this%device_comp_eig(amg, n)
450 end if
451 if (present(zero_init)) then
452 zero_initial_guess = zero_init
453 else
454 zero_initial_guess = .false.
455 end if
456 max_iter = this%max_iter
457
458 associate( w_d => this%w_d, r_d => this%r_d, d_d => this%d_d )
459
460 if (.not. zero_initial_guess) then
461 call amg%device_matvec(this%w, x, w_d, x_d, this%lvl)
462 end if
463
464 thet = this%tha
465 delt = this%dlt
466 s1 = thet / delt
467 rhok = 1.0_rp / s1
468
469 ! First iteration
470 tmp1 = 1.0_rp / thet
471 call amg_device_cheby_solve_part1(r_d, f_d, w_d, x_d, d_d, &
472 tmp1, n, zero_initial_guess)
473 ! Rest of iterations
474 do iter = 2, max_iter
475 call amg%device_matvec(this%w, this%d, w_d, d_d, this%lvl)
476
477 rhokp1 = 1.0_rp / (2.0_rp * s1 - rhok)
478 tmp1 = rhokp1 * rhok
479 tmp2 = 2.0_rp * rhokp1 / delt
480 rhok = rhokp1
481
482 call amg_device_cheby_solve_part2(r_d, w_d, d_d, x_d, tmp1, tmp2, n)
483
484 end do
485 end associate
486 end subroutine amg_device_cheby_solve
487
493 subroutine amg_jacobi_init(this, n, lvl, max_iter)
494 class(amg_jacobi_t), intent(inout), target :: this
495 integer, intent(in) :: n
496 integer, intent(in) :: lvl
497 integer, intent(in) :: max_iter
498
499 allocate(this%d(n))
500 allocate(this%w(n))
501 allocate(this%r(n))
502 this%n = n
503 this%lvl = lvl
504 this%max_iter = max_iter
505 this%omega = 0.7_rp
506
507 end subroutine amg_jacobi_init
508
510 subroutine amg_jacobi_free(this)
511 class(amg_jacobi_t), intent(inout), target :: this
512 if (allocated(this%d)) then
513 deallocate(this%d)
514 end if
515 if (allocated(this%w)) then
516 deallocate(this%w)
517 end if
518 if (allocated(this%r)) then
519 deallocate(this%r)
520 end if
521 end subroutine amg_jacobi_free
522
526 subroutine amg_jacobi_diag(this, amg, n)
527 class(amg_jacobi_t), intent(inout) :: this
528 type(tamg_hierarchy_t), intent(inout) :: amg
529 integer, intent(in) :: n
530 real(kind=rp) :: val
531 integer :: i
532 do i = 1, n
533 call tamg_sample_matrix_val(val, amg, this%lvl, i, i)
534 this%d(i) = 1.0_rp / val
535 end do
536 this%recompute_diag = .false.
537 end subroutine amg_jacobi_diag
538
544 subroutine amg_jacobi_solve(this, x, f, n, amg, niter)
545 class(amg_jacobi_t), intent(inout) :: this
546 integer, intent(in) :: n
547 real(kind=rp), dimension(n), intent(inout) :: x
548 real(kind=rp), dimension(n), intent(inout) :: f
549 class(tamg_hierarchy_t), intent(inout) :: amg
550 type(ksp_monitor_t) :: ksp_results
551 integer, optional, intent(in) :: niter
552 integer :: iter, max_iter
553 real(kind=rp) :: rtr, rnorm
554 integer :: i
555
556 if (this%recompute_diag) then
557 call this%comp_diag(amg, n)
558 end if
559
560 if (present(niter)) then
561 max_iter = niter
562 else
563 max_iter = this%max_iter
564 end if
565
566 ! x = x + omega * Dinv( f - Ax )
567 associate( w => this%w, r => this%r, d => this%d)
568 do iter = 1, max_iter
569 w = 0.0_rp
571 call amg%matvec(w, x, this%lvl)
572 !$omp parallel private(i)
574 !$omp do
575 do i = 1, n
576 r(i) = f(i) - w(i)
577 end do
578 !$omp end do
580 !$omp do
581 do i = 1, n
582 r(i) = r(i) * d(i)
583 end do
584 !$omp end do
586 !$omp do
587 do i = 1, n
588 x(i) = x(i) + this%omega * r(i)
589 end do
590 !$omp end do
591 !$omp end parallel
592 end do
593 end associate
594 end subroutine amg_jacobi_solve
595
596 subroutine amg_smoo_monitor(lvl, smoo)
597 integer, intent(in) :: lvl
598 class(amg_cheby_t), intent(in) :: smoo
599 character(len=LOG_SIZE) :: log_buf
600
601 write(log_buf, '(A8,I2,A28)') '-- level', lvl, '-- init smoother: Chebyshev'
602 call neko_log%message(log_buf)
603 write(log_buf, '(A22,I6)') 'Iterations:', smoo%max_iter
604 call neko_log%message(log_buf)
605 end subroutine amg_smoo_monitor
606
607 subroutine amg_cheby_monitor(lvl, lam)
608 integer, intent(in) :: lvl
609 real(kind=rp), intent(in) :: lam
610 character(len=LOG_SIZE) :: log_buf
611
612 write(log_buf, '(A12,I2,A29,F12.3)') '-- AMG level', lvl, &
613 '-- Chebyshev approx. max eig', lam
614 call neko_log%message(log_buf)
615 end subroutine amg_cheby_monitor
616
617end module tree_amg_smoother
__device__ T solve(const T u, const T y, const T guess, const T nu, const T kappa, const T B)
Map a Fortran array to a device (allocate and associate)
Definition device.F90:83
Copy data between host and device (or device and device)
Definition device.F90:72
Unmap a Fortran array from a device (deassociate and free)
Definition device.F90:89
subroutine, public device_add2(a_d, b_d, n, strm)
Vector addition .
subroutine, public device_add3s2(a_d, b_d, c_d, c1, c2, n, strm)
Returns .
subroutine, public device_rzero(a_d, n, strm)
Zero a real vector.
subroutine, public device_sub2(a_d, b_d, n, strm)
Vector substraction .
subroutine, public device_copy(a_d, b_d, n, strm)
Copy a vector .
real(kind=rp) function, public device_glsc3(a_d, b_d, c_d, n, strm)
Weighted inner product .
real(kind=rp) function, public device_glsc2(a_d, b_d, n, strm)
Weighted inner product .
subroutine, public device_cmult2(a_d, b_d, c, n, strm)
Multiplication by constant c .
Implements device kernels for use with TreeAMG smoothers.
subroutine, public amg_device_cheby_solve_part1(r_d, f_d, w_d, x_d, d_d, inv_thet, n, zero_initial)
subroutine, public amg_device_cheby_solve_part2(r_d, w_d, d_d, x_d, tmp1, tmp2, n)
Device abstraction, common interface for various accelerators.
Definition device.F90:34
integer, parameter, public host_to_device
Definition device.F90:48
Gather-scatter.
Implements the base abstract type for Krylov solvers plus helper types.
Definition krylov.f90:34
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 cmult(a, c, n)
Multiplication by constant c .
Definition math.f90:507
subroutine, public cmult2(a, b, c, n)
Multiplication by constant c .
Definition math.f90:522
real(kind=rp) function, public glsc3(a, b, c, n)
Weighted inner product .
Definition math.f90:1290
real(kind=rp) function, public glsc2(a, b, n)
Weighted inner product .
Definition math.f90:1269
subroutine, public add2(a, b, n)
Vector addition .
Definition math.f90:903
subroutine, public add3s2(a, b, c, c1, c2, n)
Returns .
Definition math.f90:1096
subroutine, public col2(a, b, n)
Vector multiplication .
Definition math.f90:1049
subroutine, public copy(a, b, n)
Copy a vector .
Definition math.f90:294
subroutine, public sub2(a, b, n)
Vector substraction .
Definition math.f90:951
subroutine, public add2s2(a, b, c1, n)
Vector addition with scalar multiplication (multiplication on second argument)
Definition math.f90:1001
Build configurations.
integer, parameter neko_bcknd_device
integer, parameter, public rp
Global precision used in computations.
Definition num_types.f90:14
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
Implements scalar_projector_t.
Implements smoothers for use with TreeAMG matrix vector product.
subroutine amg_cheby_free(this)
free cheby data
subroutine amg_jacobi_init(this, n, lvl, max_iter)
Initialization of Jacobi (this is expensive...)
subroutine amg_cheby_power(this, amg, n)
Power method to approximate largest eigenvalue.
subroutine amg_device_cheby_power(this, amg, n)
Power method to approximate largest eigenvalue.
subroutine amg_cheby_solve(this, x, f, n, amg, zero_init)
Chebyshev smoother From Saad's iterative methods textbook.
subroutine amg_cheby_init(this, n, lvl, max_iter)
Initialization of chebyshev.
subroutine amg_jacobi_free(this)
free jacobi data
subroutine amg_jacobi_diag(this, amg, n)
SAMPLE MATRIX DIAGONAL VALUES (DO NOT USE, EXPENSIVE)
subroutine amg_device_cheby_solve(this, x, f, x_d, f_d, n, amg, zero_init)
Chebyshev smoother From Saad's iterative methods textbook.
subroutine amg_jacobi_solve(this, x, f, n, amg, niter)
Jacobi smoother.
subroutine amg_smoo_monitor(lvl, smoo)
subroutine amg_cheby_monitor(lvl, lam)
Implements utilities for the TreeAMG hierarchy structure.
subroutine, public tamg_sample_matrix_val(val, amg, lvl, i, j)
Sample the values in a matix (expensive, use with caution)
Implements the base type for TreeAMG hierarchy structure.
Definition tree_amg.f90:34
Gather-scatter kernel.
Type for storing initial and final residuals in a Krylov solver.
Definition krylov.f90:57
Projector for scalar boundary conditions.
Type for a TreeAMG hierarchy.
Definition tree_amg.f90:101
Type for Chebyshev iteration using TreeAMG matvec.
Type for Chebyshev iteration using TreeAMG matvec.