compiler: Introducing min/max bounds to replace 'bf' elemental functions #1673

georgebisbas · 2021-04-20T15:43:35Z

This PR is removing Elemental Functions from Devito generated code. The so-called "bf" functions.
The most important part for review is in devito/passes/iet/misc.py.

HaloSpot is now correctly propagated in the IET. Previously it was dropped to the floor when existing in between outer dimensions. This was impacting the number of collapsed loops. Issue compiler: buggy compose_nodes, drops HaloSpots #1723

So now you get:

  {
    /* Begin section0 */
    START_TIMER(section0)
    if (x0_blk0_size == 0 || y0_blk0_size == 0)
    {
      return;
    }
    #pragma omp parallel num_threads(nthreads)
    {
      #pragma omp for collapse(2) schedule(dynamic,1)
      for (int x0_blk0 = x_m; x0_blk0 <= x_M; x0_blk0 += x0_blk0_size)
      {
        for (int y0_blk0 = y_m; y0_blk0 <= y_M; y0_blk0 += y0_blk0_size)
        {
          for (int x = x0_blk0; x <= MIN(x_M, x0_blk0 + x0_blk0_size - 1); x += 1)
          {
            for (int y = y0_blk0; y <= MIN(y_M, y0_blk0 + y0_blk0_size - 1); y += 1)
            {
              #pragma omp simd aligned(damp,u,vp:32)
              for (int z = z_m; z <= z_M; z += 1)
              {
                float r10 = 1.0F/dt;
                float r9 = 1.0F/(dt*dt);
                float r8 = 1.0F/(vp[x + 4][y + 4][z + 4]*vp[x + 4][y + 4][z + 4]);
                u[t2][x + 4][y + 4][z + 4] = (r10*(damp[x + 1][y + 1][z + 1]*u[t0][x + 4][y + 4][z + 4]) + r8*(-r9*(-2>
              }
            }
          }
        }
      }
    }
    STOP_TIMER(section0,timers)

instead of:

  for (int time = time_m, t0 = (time)%(3), t1 = (time + 2)%(3), t2 = (time + 1)%(3); time <= time_M; time += 1, t0 = (>
  {
    /* Begin section0 */
    START_TIMER(section0)
    bf0(damp_vec,dt,u_vec,vp_vec,t0,t1,t2,x0_blk0_size,x_M - (x_M - x_m + 1)%(x0_blk0_size),x_m,y0_blk0_size,y_M - (y_>
    bf0(damp_vec,dt,u_vec,vp_vec,t0,t1,t2,x0_blk0_size,x_M - (x_M - x_m + 1)%(x0_blk0_size),x_m,(y_M - y_m + 1)%(y0_bl>
    bf0(damp_vec,dt,u_vec,vp_vec,t0,t1,t2,(x_M - x_m + 1)%(x0_blk0_size),x_M,x_M - (x_M - x_m + 1)%(x0_blk0_size) + 1,>
    bf0(damp_vec,dt,u_vec,vp_vec,t0,t1,t2,(x_M - x_m + 1)%(x0_blk0_size),x_M,x_M - (x_M - x_m + 1)%(x0_blk0_size) + 1,>
    STOP_TIMER(section0,timers)
    /* End section0 */

where

bf:...

void bf0(struct dataobj *restrict damp_vec, const float dt, struct dataobj *restrict u_vec, struct dataobj *restrict v>
{
  float (*restrict damp)[damp_vec->size[1]][damp_vec->size[2]] __attribute__ ((aligned (64))) = (float (*)[damp_vec->s>
  float (*restrict u)[u_vec->size[1]][u_vec->size[2]][u_vec->size[3]] __attribute__ ((aligned (64))) = (float (*)[u_ve>
  float (*restrict vp)[vp_vec->size[1]][vp_vec->size[2]] __attribute__ ((aligned (64))) = (float (*)[vp_vec->size[1]][>

  if (x0_blk0_size == 0 || y0_blk0_size == 0)
  {
    return;
  }
  #pragma omp parallel num_threads(nthreads)
  {
    #pragma omp for collapse(2) schedule(dynamic,1)
    for (int x0_blk0 = x_m; x0_blk0 <= x_M; x0_blk0 += x0_blk0_size)
    {
      for (int y0_blk0 = y_m; y0_blk0 <= y_M; y0_blk0 += y0_blk0_size)
      {
        for (int x = x0_blk0; x <= x0_blk0 + x0_blk0_size - 1; x += 1)
        {
          for (int y = y0_blk0; y <= y0_blk0 + y0_blk0_size - 1; y += 1)
          {
            #pragma omp simd aligned(damp,u,vp:32)
            for (int z = z_m; z <= z_M; z += 1)
            {
              float r10 = 1.0F/dt;
              float r9 = 1.0F/(dt*dt);
              float r8 = 1.0F/(vp[x + 4][y + 4][z + 4]*vp[x + 4][y + 4][z + 4]);
              u[t2][x + 4][y + 4][z + 4] = (r10*(damp[x + 1][y + 1][z + 1]*u[t0][x + 4][y + 4][z + 4]) + r8*(-r9*(-2.0>
            }

codecov · 2021-04-20T16:00:12Z

Codecov Report

Merging #1673 (6f243ac) into master (1471d53) will decrease coverage by 1.68%.
The diff coverage is 96.89%.

❗ Current head 6f243ac differs from pull request most recent head c38526b. Consider uploading reports for the commit c38526b to get more accurate results

@@            Coverage Diff             @@
##           master    #1673      +/-   ##
==========================================
- Coverage   89.60%   87.91%   -1.69%     
==========================================
  Files         206      201       -5     
  Lines       33163    32669     -494     
  Branches     4316     4281      -35     
==========================================
- Hits        29716    28722     -994     
- Misses       2958     3443     +485     
- Partials      489      504      +15

Impacted Files	Coverage Δ
devito/core/arm.py	`48.14% <0.00%> (ø)`
devito/ir/iet/visitors.py	`80.71% <ø> (-0.19%)`	⬇️
tests/test_dse.py	`94.51% <93.02%> (-5.29%)`	⬇️
devito/passes/iet/misc.py	`86.15% <95.45%> (-5.85%)`	⬇️
tests/conftest.py	`90.54% <96.77%> (-0.92%)`	⬇️
devito/core/autotuning.py	`91.15% <100.00%> (-2.04%)`	⬇️
devito/core/cpu.py	`100.00% <100.00%> (ø)`
devito/core/gpu.py	`95.69% <100.00%> (+0.56%)`	⬆️
devito/ir/iet/utils.py	`80.95% <100.00%> (+5.51%)`	⬆️
devito/passes/iet/parpragma.py	`92.10% <100.00%> (-0.08%)`	⬇️
... and 45 more

Continue to review full report at Codecov.

Legend - Click here to learn more
Δ = absolute <relative> (impact), ø = not affected, ? = missing data
Powered by Codecov. Last update 1471d53...c38526b. Read the comment docs.

review-notebook-app · 2021-04-21T09:45:21Z

Check out this pull request on

See visual diffs & provide feedback on Jupyter Notebooks.

Powered by ReviewNB

tests/test_autotuner.py

FabioLuporini

it's not easy to follow the implementation of relax_incr_dimension

You could try improving the comments, maybe you could also add a couple of examples inline just like I often do for ease of explanation

devito/core/autotuning.py

devito/ir/iet/utils.py

devito/passes/iet/misc.py

devito/core/autotuning.py

devito/core/gpu.py

devito/operator/operator.py

devito/passes/iet/misc.py

examples/cfd/01_convection_revisited.ipynb

tests/test_mpi.py

tests/test_autotuner.py

FabioLuporini · 2021-07-20T07:29:39Z

tests/conftest.py

+def assert_structure(operator, exp_trees=None, exp_iters=None):
+    """
+    Utility function that helps to check loop structure of IETs. Retrieves trees from an
+    Operator and check that blocking structure is as expected. Trees and Iterations are


... that the blocking structure...

FabioLuporini · 2021-07-20T07:30:18Z

tests/conftest.py

+
+    Example:
+
+    To check the following structure:


"To check that an Iteration tree as the following structure"

ok, (*has the following.. )?

FabioLuporini · 2021-07-20T07:30:44Z

tests/conftest.py

+
+    we call:
+
+    `trees, iters = assert_structure(op, ['t,x,y', 't,f,y'], 't,x,y,f,y')`


nitpicking: use should use ..code-block again, no?

no, AFAI saw from other examples....did some more changes to be homogeneous with the codebase, let me know if yout think it is better or not now...

ah, yes, I thought you were saying that we need to have an ending delimeter such as

code-block
code
code-block end
ok, all clear now

tests/conftest.py

FabioLuporini · 2021-07-20T07:31:55Z

tests/conftest.py

+
+    we call:
+
+    `trees, bns = get_blocked_nests(op, {'x0_blk0', 'x1_blk0'})`


same as before, use code-block ?

tests/conftest.py

FabioLuporini · 2021-07-20T07:39:13Z

tests/test_dimension.py

@@ -1072,11 +1072,12 @@ def test_no_fusion_simple(self):
                Eq(g, f + 1, implicit_dims=[ctime])]

        op = Operator(eqns)
-        exprs = FindNodes(Expression).visit(op._func_table['bf0'].root)
+        _, bns = get_blocked_nests(op, {'x0_blk0', 'x1_blk0'})


imho all these calls to assert_blocking want blank lines around

So, not as before right?
i.e. we add two empty lines ?

FabioLuporini · 2021-07-20T07:40:03Z

tests/test_dle.py

-    assert tree[3].dim.is_Incr and tree[3].dim.parent is tree[1].dim and\
-        tree[3].dim.root is y
-    assert not tree[4].dim.is_Incr and tree[4].dim is zi and tree[4].dim.parent is z
+    _, _ = assert_structure(op, ['t,i0x0_blk0,i0y0_blk0,i0x,i0y,i0z'])


I don't know, you tell me?

tests/test_dse.py

mloubout

Some comment, very nice!

devito/passes/iet/parpragma.py

examples/cfd/01_convection_revisited.ipynb

tests/conftest.py

tests/test_dimension.py

mloubout · 2021-07-20T15:29:59Z

tests/test_dle.py

+                            opt=('blocking', {'openmp': True,
+                                              'blockinner': blockinner,
+                                              'par-collapse-ncores': 1}))
+    for opi in [op, op2]:


why not parametize over openmp?

This will lead to redundantly run additional tests as there is CI for openmp and no-openmp, no?

I don't see how, it won't change anything only split this test in two cases rather than having them as if/else inside the test.

Ah yes, you are right, pushed a change, let me know if you like it.

tests/test_skewing.py

georgebisbas

Thanks for the review!

devito/passes/iet/parpragma.py

examples/cfd/01_convection_revisited.ipynb

tests/conftest.py

tests/test_dimension.py

georgebisbas · 2021-07-21T12:30:42Z

tests/test_dle.py

+                            opt=('blocking', {'openmp': True,
+                                              'blockinner': blockinner,
+                                              'par-collapse-ncores': 1}))
+    for opi in [op, op2]:


This will lead to redundantly run additional tests as there is CI for openmp and no-openmp, no?

tests/test_skewing.py

FabioLuporini

we nearly ready to land!

tests/conftest.py

FabioLuporini · 2021-07-23T07:36:02Z

tests/conftest.py

+
+    if exp_trees is not None:
+        exp_trees = [i.replace(',', '') for i in exp_trees]  # 't,x,y' -> 'txy'
+        tree_struc = (["".join(mapper.get(i.dim.name, i.dim.name) for i in j)


comment about what this line's doing? just like the one in the line above

tests/conftest.py

FabioLuporini · 2021-07-23T07:39:10Z

tests/conftest.py

+    trees = retrieve_iteration_tree(operator)
+    for tree in trees:
+        iterations = [i for i in tree if i.dim.is_Incr]  # Collect Incr dimensions
+        parallel_blocks = FindNodes(ParallelBlock).visit(tree)


You only need this tree traversal if you get inside the if iterations branch right? so move it below inside the try?

FabioLuporini · 2021-07-23T07:39:46Z

tests/test_autotuner.py

@@ -238,8 +238,7 @@ def test_multiple_blocking():
                  opt=('blocking', {'openmp': False}))

    # First of all, make sure there are indeed two different loop nests
-    assert 'bf0' in op._func_table
-    assert 'bf1' in op._func_table
+    _, _ = assert_blocking(op, {'x0_blk0', 'x1_blk0'})


just assert_blocking(op, {'x0_blk0', 'x1_blk0'}) no need to catch the return value if ever used!

replicate this change wherever needed

tests/test_dse.py

FabioLuporini · 2021-07-26T10:37:17Z

tests/conftest.py

+    """
+    mapper = {'time': 't'}
+    trees = retrieve_iteration_tree(operator)
+    iters = FindNodes(Iteration).visit(operator)


this is not needed unless exp_iters is not None

There are cases where the structure is simple, so exp_iters is not tested as redundant. There are also tests that use Iterations later. So it is needed.

FabioLuporini · 2021-07-26T10:37:48Z

tests/conftest.py

+            for f
+                for y
+
+    we call(Note: `time` mapped to `t`):


space, or rather change the for t into for time in the code-block above?

FabioLuporini

I left a couple of tiny, last-minute comments, but to be fair I'm fine if they're addressed in the next PR. Approving -- well done! This is a really nice piece of work

mloubout · 2021-07-26T13:04:30Z

tests/test_dle.py

-    assert len(trees) == 1
-    tree = trees[0]
-    assert len(tree) == exp_iters
+                           opt=('blocking', {'openmp': True, 'blockinner': blockinner,


openmp: openmp?

nitpicking: blockinner could probably be moved to parametrize as well

WDYM, it is alreday there?

FabioLuporini · 2021-07-26T14:01:44Z

tests/conftest.py

        exp_iters = exp_iters.replace(',', '')  # 't,x,y' -> 'txy'
        iter_struc = "".join(mapper.get(i.dim.name, i.dim.name) for i in iters)
        assert iter_struc == exp_iters

-    return trees, iters
+    return


georgebisbas added compiler RFC testing WIP Still work in progress labels Apr 21, 2021

georgebisbas self-assigned this Apr 21, 2021

georgebisbas commented Apr 21, 2021

View reviewed changes

tests/test_autotuner.py Outdated Show resolved Hide resolved

georgebisbas added this to In progress in Temporal blocking in Devito Apr 21, 2021

georgebisbas mentioned this pull request Apr 21, 2021

TEST ONLY #1675

Closed

georgebisbas force-pushed the minmaxbounds branch 2 times, most recently from 0823044 to 754e1a5 Compare April 26, 2021 15:44

FabioLuporini requested changes Apr 28, 2021

View reviewed changes

mloubout reviewed Apr 28, 2021

View reviewed changes

devito/passes/iet/misc.py Outdated Show resolved Hide resolved

FabioLuporini reviewed Apr 30, 2021

View reviewed changes

FabioLuporini removed the RFC label Apr 30, 2021

georgebisbas force-pushed the minmaxbounds branch from 6aee64a to 41fae93 Compare May 5, 2021 13:44

georgebisbas force-pushed the minmaxbounds branch 2 times, most recently from 63b2074 to 6bd01b0 Compare June 1, 2021 12:52

georgebisbas requested review from FabioLuporini and mloubout June 3, 2021 15:45

FabioLuporini requested changes Jun 10, 2021

View reviewed changes

georgebisbas force-pushed the minmaxbounds branch 3 times, most recently from f8b020f to b75e1ca Compare June 11, 2021 10:07

georgebisbas force-pushed the minmaxbounds branch from 1e21971 to 13c79c1 Compare June 18, 2021 07:38

georgebisbas commented Jun 18, 2021

View reviewed changes

tests/test_autotuner.py Outdated Show resolved Hide resolved

georgebisbas force-pushed the minmaxbounds branch 2 times, most recently from 8254b84 to 8844b0a Compare June 28, 2021 09:33

georgebisbas changed the title ~~[RFC] compiler: Introducing min/max bounds to replace 'bf' elemental functions~~ compiler: Introducing min/max bounds to replace 'bf' elemental functions Jun 28, 2021

georgebisbas added 4 commits July 19, 2021 13:42

compiler: simplify algorithm and docs in relax

0ce6a13

tests: strengthen testing

097f08c

tests: update tests with helper funcs

3b32223

tests: update tests with helper funcs II

df1e67d

georgebisbas force-pushed the minmaxbounds branch from a2f0416 to df1e67d Compare July 19, 2021 10:44

FabioLuporini reviewed Jul 20, 2021

View reviewed changes

tests: simplify by ret ParallelBlocks

29b832b

georgebisbas force-pushed the minmaxbounds branch from 4252061 to 29b832b Compare July 20, 2021 15:12

mloubout reviewed Jul 20, 2021

View reviewed changes

georgebisbas commented Jul 21, 2021

View reviewed changes

tests: improve docs, parametrize

9cbad5f

georgebisbas requested review from FabioLuporini and mloubout July 21, 2021 14:49

FabioLuporini reviewed Jul 23, 2021

View reviewed changes

tests: improve tests, docs

4d25407

georgebisbas force-pushed the minmaxbounds branch from 30cdbc5 to 4d25407 Compare July 26, 2021 08:13

georgebisbas requested a review from FabioLuporini July 26, 2021 08:30

georgebisbas removed the WIP Still work in progress label Jul 26, 2021

FabioLuporini reviewed Jul 26, 2021

View reviewed changes

FabioLuporini approved these changes Jul 26, 2021

View reviewed changes

mloubout reviewed Jul 26, 2021

View reviewed changes

mloubout approved these changes Jul 26, 2021

View reviewed changes

FabioLuporini reviewed Jul 26, 2021

View reviewed changes

tests: improve conftest, dle

c38526b

georgebisbas force-pushed the minmaxbounds branch from c7735d0 to c38526b Compare July 26, 2021 14:10

FabioLuporini merged commit 4f2a1fb into master Jul 26, 2021

FabioLuporini deleted the minmaxbounds branch July 26, 2021 15:29

georgebisbas moved this from In progress to Done in Temporal blocking in Devito Jul 27, 2021

FabioLuporini mentioned this pull request Jul 29, 2021

add guard for empty regions in mpi=full #1314

Closed


		we call:

		`trees, iters = assert_structure(op, ['t,x,y', 't,f,y'], 't,x,y,f,y')`


		we call:

		`trees, bns = get_blocked_nests(op, {'x0_blk0', 'x1_blk0'})`

compiler: Introducing min/max bounds to replace 'bf' elemental functions #1673

compiler: Introducing min/max bounds to replace 'bf' elemental functions #1673

Conversation

georgebisbas commented Apr 20, 2021 • edited

codecov bot commented Apr 20, 2021 • edited

Codecov Report

review-notebook-app bot commented Apr 21, 2021

FabioLuporini left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

mloubout left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

georgebisbas left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

FabioLuporini left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

FabioLuporini left a comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

georgebisbas commented Apr 20, 2021 •

edited

codecov bot commented Apr 20, 2021 •

edited