intel / yask Goto Github PK

Hello Charles,

I am very interested in this work and I am learning about this framework. I try adding a new stencil into this tool.
I have a question about how to decide the input data. I check the existed stencils, such as AveStencile and Iso3dfdStencil, but I still have no idea about the location of input data.

Thanks for your help.

Seems mpi=1 is required even for single node builds.

icpc -V
Intel(R) C++ Intel(R) 64 Compiler for applications running on Intel(R) 64, Version 17.0.0.098 Build 20160721

make arch=hsw stencil=iso3dfd
Header files generated.
icpc -g -O3 -std=c++11 -Wall -xCORE-AVX2 -debug extended -Fa -restrict -ansi-alias -fno-alias -fimf-precision=low -fast-transcendentals -no-prec-sqrt -no-prec-div -fp-model fast=2 -fno-protect-parens -rcd -ftz -fma -fimf-domain-exclusion=none -qopt-assume-safe-padding -qopt-report=5 -qopt-report-phase=VEC,PAR,OPENMP,IPO,LOOP -no-diag-message-catalog -DMAX_EXCH_DIST=1 -DUSE_INTRIN256 -DREAL_BYTES=4 -DLAYOUT_3D=Layout_123 -DLAYOUT_4D=Layout_2314 -DTIME_DIM_SIZE=2 -DDEF_RANK_SIZE=1024 -DDEF_BLOCK_SIZE=64 -DDEF_BLOCK_THREADS=2 -DDEF_THREAD_FACTOR=1 -DDEF_PAD=1 -DARCH_HSW -DNO_STORE_INTRINSICS -DUSE_STREAMING_STORE -fopenmp -c -o src/stencil_main.hsw.o src/stencil_main.cpp
icpc: remark #10397: optimization reports are generated in *.optrpt files in the output location
icpc -g -O3 -std=c++11 -Wall -xCORE-AVX2 -debug extended -Fa -restrict -ansi-alias -fno-alias -fimf-precision=low -fast-transcendentals -no-prec-sqrt -no-prec-div -fp-model fast=2 -fno-protect-parens -rcd -ftz -fma -fimf-domain-exclusion=none -qopt-assume-safe-padding -qopt-report=5 -qopt-report-phase=VEC,PAR,OPENMP,IPO,LOOP -no-diag-message-catalog -DMAX_EXCH_DIST=1 -DUSE_INTRIN256 -DREAL_BYTES=4 -DLAYOUT_3D=Layout_123 -DLAYOUT_4D=Layout_2314 -DTIME_DIM_SIZE=2 -DDEF_RANK_SIZE=1024 -DDEF_BLOCK_SIZE=64 -DDEF_BLOCK_THREADS=2 -DDEF_THREAD_FACTOR=1 -DDEF_PAD=1 -DARCH_HSW -DNO_STORE_INTRINSICS -DUSE_STREAMING_STORE -fopenmp -c -o src/stencil_calc.hsw.o src/stencil_calc.cpp
icpc: remark #10397: optimization reports are generated in *.optrpt files in the output location
src/stencil_calc.cpp(701): error: identifier "MPI_INTEGER8" is undefined
MPI_Bcast(&coords[rn][0], num_dims, MPI_INTEGER8,
^

src/stencil_calc.cpp(701): error: identifier "MPI_Bcast" is undefined
MPI_Bcast(&coords[rn][0], num_dims, MPI_INTEGER8,
^

compilation aborted for src/stencil_calc.cpp (code 2)
make: *** [src/stencil_calc.hsw.o] Error 2

no need to sync between eq groups that aren't dependent on each other.

Currently, the dependency-calculation code in the compiler is pretty primitive. It is especially poor between sub-domains.

Look at stencil to see if it needs to exchange halo along edges or at corners, etc.

Example:

• make clean; make -j stencil=3axis arch=snb
• bin/yask.sh -ranks 2 -stencil 3axis -arch snb -bad-option

Can only see the following:

YASK Kernel: .
application called MPI_Abort(MPI_COMM_WORLD, 1) - process 1
YASK Kernel: .
application called MPI_Abort(MPI_COMM_WORLD, 1) - process 0

Just an alias to call free function yask_get_version_string().

Non-canonical implies dependencies don't have to be in step dim.
API would be something like run_stencil(stencil_name, slice, steps)

Need to break into rectilinear sub-sub-domains or devise some other scheme to vectorize as much as possible.

Hello,

It fails when I try using INIT_GRID_2D to design a kernel, but I find the INIT_GRID_2D method defined in src/foldBuilder/Expr.hpp. Is there any way to use INIT_GRID_2D? Thanks for your help.

Suggest using regular-expressions instead of simple sub-string matches.

Sorry for the pettiness of this issue, but I believe that good descriptions are helpful: the first paragraph of the README reads: "YASK--Yet Another Stencil Kernel: A framework to facilitate exploration of the HPC stencil-performance design space, including any optimizations such as...". Should it not be, instead "including many optimizations" ?

I hope this is useful. Thank you for your wonderful work!

An existing code sequence like
cerr << "Error: run_solution() called without calling prepare_solution() first.\n"; exit_yask(1);
Should throw an exception with the formatted string as a payload.
In the kernel code, most of the code to be replaced should be in this cerr...exit_yask(1) form.
In the compiler code, most of the code to be replaced should be in the form cerr...exit(1).
There could be different types of exceptions: parameter-out-of-range, invalid-state, etc., or initially we could just have one generic yask exception.
The program in kernel/yask_main.cpp should catch the exception and call something similar to the existing exit_yask(). Same with compiler/main.cpp.
The fix should include some tests in the C++ and perl API test programs.

Like v = element_read(x); v += n; element_write(v, x);

Performance is reduced by several factors when the domain size is not a multiple of the vector-cluster size in each dim. Need to add remainder loops to handle partial vectors.

First solution is ok. Subsequent ones have fewer threads.

location: src/kernel/lib/realv_grids.cpp:void RealVecGridBase::share_storage(yk_grid_ptr source)

in the implementation, get_num_dims() may return 4 if the time dimension is present, but then get_rank_domain_size(dname) is called, which will fail if dname == 't' (as t isn't a domain dimension)

Automatically distribute size among ranks as a default.
For strong and weak-scaling, provide a heuristic to set rank topology.

Hi,

I am evaluating your framework on a machine with a socketed Xeon Phi 7210F and the MCDRAM configured in the default cache mode. I compiled the 3axis stencil by using the following command:

make clean; make stencil=3axis arch=knl

If I run the stencil using yask.sh, I get the following results:

bin/yask.sh -stencil 3axis -arch knl -d 512 -t 1:

best-throughput (est-FLOPS):       138.865G

However, running the same stencil directly using the executable with the same settings results in significantly lower performance:

bin/yask_kernel.3axis.knl.exe -d 512 -t 1:

best-throughput (est-FLOPS):       31.4167G

Is there a specific reason for this? Should I always run from the sh file?

Has been steadily declining. Used to be around 18 Gpts/s. Now at 15.

Currently ignores sub-domain of dependent(s).

Test with MPI on z-axis of awp_elastic stencil

Settings can now be made via apply_command_line_options().

Simply first domain minus halo and last domain plus halo.
Use GET/SET_GRID_API macros in realv_grids.cpp.

I am trying to evaluate the performance of the 3axis stencil in your framework with different radius values on a Xeon Phi 7210F and the MCDRAM configured in the default cache mode flat mode. My compiler is ICC 2018.1. I am trying to improve the performance of this stencil using temporal blocking (-r and -rt switches), since the general consensus is that this stencil is memory-bound, especially at low-order, and could benefit from temporal blocking. However, I have not been successful to improve the performance using temporal blocking, unless when the input size is small (smaller than 128^3). Should I at all expect performance improvement for this stencil when large inputs are used (512^3 and above) using temporal blocking, or is the ~140 GFLOPs that can be achieved using the default settings, the maximum achievable performance?

Hello,

It fails when I try adding an IF condition in the kernel, such as grid(t+1, x, y, z) IS_EQUIV_TO v IF (v == constNum(1.0)). However, it works about grid(t+1, x, y, z) IS_EQUIV_TO v IF (z == last_index(z)). I think both IF conditions return the same type. Do you have any ideas about this problem? Thank you so much.

E.g., run in a max size region.

Template parameter should specify size.

E.g., M[x,y,z,i] can be accessed via pointer p to M[x,y,0,0] as p[z * si + i], where si is size of alloc in i dim.

it's leaving some useless files under /lib, such as devito_ctx0_yk_hook0.py, devito_ctx0_yk_hook0.pyc, ...

Rumored to have a small perf advantage.

Will require adding APIs, removing functionality, and/or moving code from kernel/lib to yask_main.cpp.
This is not critical, but it would showcase the API if we actually used it!

Can you kindly elaborate how the boundary condition for the 3axis stencil is handled? By boundary condition I mean the neighbors of cells on the grid boundaries that fall outside of the gird. From the WOLFHPC16 slides it seems that the allocated grid is (2 * radius) points bigger than the problem domain in each dimension, allowing such neighbors to be read from a valid memory address. Is this correct? If yes, am I correct to say that the input size that we set in YASK is the actual problem size, but the size of the grid that is allocated in memory is bigger due to these out-of-bound neighbors?

Overlap halo exchange with internal block computations.

Using https://github.com/LLNL/zfp or something similar.

Avoids running everything in HBM via numactl.
Needed for DUG prototype.

Temps would be similar to grids, but data would not be available outside of computation.
Stencil compiler should determine required "halo" size based on accesses in stencil.
Actual size would be based on block-size plus these halos.
Would need a set of temps for each thread team. These would be reused across blocks.

Leave specified number of steps as an option.
Can implement as 2 minimums that must both be met.

Add a test when fixed.

Should be based on L2 size and number of grids read and written per stencil group.

set_domain_dim_names(<dimensions>) should fail immediately if t is provided as domain dimension, but it doesn't

Has dropped by 4x on iso3dfd!

Hello,

I got an error " Please verify that both the operating system and the processor support Intel(R) MOVBE, F16C, FMA, BMI, LZCNT and AVX2 instructions." when I run YASK.

Following is the command:
mpirun-n 2 -ppn2 ./stencil-run.sh -arch hsw---nr 1 -nrx 2 -d 1024 -dx 512 -b 64 -bz 96

CPU : Intel(R) Xeon(R) CPU E5-2670 0 @ 2.60GHz
MIC : Intel Xeon Phi coprocessors 5110P

Thanks for your help.

Put operator+(), etc. in API.