Intel Xeon Phi

Innen: KIFÜ Wiki
A lap korábbi változatát látod, amilyen Szigeti(AT)niif.hu (vitalap | szerkesztései) 2016. május 3., 10:21-kor történt szerkesztése után volt. (Hosszredukció)

Az oldalon található információk segítséget nyújtanak a HPC felhasználóinknak, hogy alkalmazásaikat miként portolhatják valamint optimalizálhatják a hatékonyabb futást szem előtt tartva Intel Xeon Phi koprocesszor kártyákon.

Hardver bemutatása

Intel® Xeon Phi™ Coprocessor 7120P

CPU threads/Core max threads CPU Freq Memória L1 cache L2 cache Interfész Ptot Számítási kapacitás
61 4 Hardver 244 1238 GHz 16GB GDDR5 32KB/core 512KB/core PCIe 300W 1,1 PFlop/s
OpenMP-nél csak 240 ~100 cycles

~174 GB/s

~3 cycles ~10 cycles ~7 GB/s

~1 microsecond

Az Intel® 1. generációs Xeon Phi™ koprocesszorában található 61 processzor ugyanazt a 16GB memóriát éri el, de semmilyen módon se fér hozzá közvetlenül a hoszt gép fizikai memóriájához.
Budapest2 és Debrecen3-Phi gépeinkben számítási nodonként 2-2 Phi™ koprocesszor található.
A kártyákon általános célú Linux kernel fut melyre a belépés limitáltan csak a rendszergazdák számára lehetséges SSH-n keresztül a hoszt gépről. Természetesen a kártyára belépve olyan érzésünk lehet, mintha egy sok processzoros gépen dolgoznánk, ami rendelkezik rengeteg memóriával. Ez így is van, de a hatékonyabb kihasználás érdekében az SSH-n keresztüli elérés nem lehetséges, csak a Slurm segítségével valamint a programozási nyelvek által támogatott csatornákon keresztül. Ennek részletezése lentebb megtalálható.

Használati módok

A koprocesszor használatai módjait két csoportba lehet sorolni. A natív és az offload használatot.
Ezen használati módokról a lenti fejezetekben több részletet is megtudhat. Az itt megfogalmazott "támogatás" csak a mi infrastruktúránkra érvényes!
A bemutatott példák c++ nyelven vannak, de természetesen c és fortran esetén is megtalálható az ekvivalens módszer.

a) nem támogatott mód

közvetlen natív használat

A natív mód arra szolgál, hogy a MIC kártyára lefordított alkalmazásunkat felmásoljuk a kártyára, majd belépünk SSH-n a kártyára és azon futtatjuk a bináris kódunkat.
Ez a legegyszerűbb használati mód, ugyanis nincs más dolgunk, csak a -mmic kapcsolót használni a fordításnál és a fordító egy máris a MIC architektúrára kész bináris készít nekünk.
Workflow a következő lenne:

  1. Alkalmazás megírása. például hello.cc
  2. Alkalmazás fordítása
    icpc -o hello-MIC -mmic hello.cc
  3. Másolás a MIC kártyára
    scp hello-MIC mic0:
  4. belépés a kártyára
    ssh mic0
  5. Alkalmazás futtatása
    ./hello-MIC

b) támogatott, de nem preferált használat

automatikus kódátvitel

Ebben az esetben nem kell átmásolnunk a bináris alkalmazásunkat a kártyára sőt még csak be se kell lépnünk, mert ezt a két lépést egy beépített tool végzi el.
A tool eléréséhez szükséges parancs neve: micnativeloadex

Ebben az esetben a workflow a következőképpen módosul:

  1. Alkalmazás megírása. például hello.cc
  2. Alkalmazás fordítása
    icpc -o hello-MIC -mmic hello.cc
  3. A tool futásához szükséges library-k megadása
    export SINK_LD_LIBRARY_PATH=/opt/intel/composerxe/compiler/lib/mic
  4. Alkalmazás futtatása
    micnativeloadex hello-MIC
BUDAPEST2[cn10] phi (0)$ export SINK_LD_LIBRARY_PATH=/opt/intel/composerxe/compiler/lib/mic
BUDAPEST2[cn10] phi (0)$ micnativeloadex hello-MIC
Hello world! I have 244 logical cores.
BUDAPEST2[cn10] phi (0)$

c) támogatott módok

Offload

Az Offload modell esetén szükséges a forráskód módosítása, mégpedig úgy, hogy azon kódrészletet, amit a koprocesszoron szeretnénk futtatni, ki kell egészíteni egy fordítási direktívával. Az alkalmazás fordításakor elegendő csak a CPU-ra való binárist készíteni mivel a modell végzi a megfelelő forrás részlet mozgatását és futtatását.
Offload modell
Példánkban jól látható, hogy a második kiíratás már a koprocesszoron fut.

#include <stdio.h>
int main(int argc, char * argv[]) {
  printf("Hello world from host!\n");
#pragma offload target(mic)
  {
    printf("Hello world from coprocessor!\n"); fflush(stdout);
  }
  printf("Bye\n");
}

Ebben a példában az operációs rendszer választja ki, hogy melyik Phi kártyán fusson le a kódrészlet. Amennyiben mi szeretnénk explicite megmondani, hogy melyik kártyán fusson, akkor a fordítási direktívában a mic kulcsszó után kell tenni egy ":"-t és a kártya sorszámát [mi esetünkben 0 vagy 1]. Ekkor így nézne ki a direktíva: #pragma offload target(mic:1)

Amennyiben az offload szekcióban függvény hívás is szerepel, akkor szükségeltetik az függvény futásához szükséges változók adatterületének átmásolása a Phi kártyára, majd a lefutás után visszamásolni a hoszt gépre. Ezen másolást az offload_attribute push és pop paraméterei segítségével határozhatjuk meg.
A következő mátrix-vektor szorzás példánál az A mátrix globális változó, míg a b és c vektorok a függvény paraméterei.

//A változók felmásolásra kerülnek a Xeon Phi-re
#pragma offload_attribute(push, target(mic))
  double* A;
  void MatrixVector(double *b, double *c, double m, double n) {
    ...
    c[] = A[..]*b[]
  }
// felmásolt objektumok visszamásolása
#pragma offload_attribute(pop)

int main() {
  ...
#pragma offload target(mic) in(A[0:n*m])
// így is meg lehetne adni: #pragma offload target(mic:0) in(A : length(n*m))	        
  {
    MatrixVector(b, c, m, n);
  }
  ...
}

Ekkor minden memória területet 2x kell másolni, egyszer fel a kártyára, majd másodszor vissza a hoszt gépre, pedig elegendő lenne csak az eredmény vektort visszamásolni. Ezen optimalizálással egy későbbi fejezetben foglalkozunk.

Offload + OpenMP

Amennyiben nem csak 1 processzort szeretnénk használni a Phi kártyánkon és az alkalmazásunkban használt számítás megengedi a párhuzamosítást, akkor az OpenMP segítségével párhuzamosíthatjuk a megfelelő számítást.
Offload modell és OpenMP
Példánk maradjon ugyanúgy a mátrix-vektor szorzás. Amennyiben a teljes kártya adta kapacitást ki szeretnénk használni, akkor az OMP threadek számát 240-re kell beállítani. Azért nem 244-re, mivel 1 CPU-t dedikáltan csak a kommunikációra használ a rendszer és 1 CPU 4 hardver threaddel rendelkezik. Ekkor a példánk így változik:

#include <omp.h>

#pragma offload_attribute(push, target(mic))
void multiply(int n, int m, double* A, double* b, double* c){
  const int nCPUs = omp_get_max_threads();   //240
  omp_set_num_threads(nCPUs);
#pragma omp parallel for schedule(dynamic, 1)
  for ( int i = 0 ; i < m ; i++)
     for ( int j = 0 ; j < n ; j++)
       c[i] += A[i*n+j] * b[j];
}
#pragma offload_attribute(pop)

int main(){
  double * A = (double*) malloc(sizeof(double)*n*m);
  ...
#pragma offload target(mic) in(A:length(n*m))
     multiply(n, m, A, b, c);


MPI + Offload + OpenMP

Abban az esetben, ha 1 számítási node nem elég, MPI segítségével össze tudunk kötni több nodet, heterogén rendszert alkotva, és azokon alkalmazható a fenti Offload + OpenMP modell.
MPI elosztás valamint Offload modell és OpenMP

Alkalmazások optimalizálása a kártyára

Ebben a fejezetben optimalizálási lehetőségeket mutatunk be, melyek segítségül szolgálhatnak az egyes alkalmazások hatékonyabb futása érdekében ezzel jobban kihasználni a rendelkezésre álló erőforrások adta számítási kapacitás.

Hosszredukció

Vektorizáció

SIMD

AoS -> SoA

Hatékony hossz választás

"Héj ciklusok" kiköszöbölése

Regularizáció

Párhuzamosítás

Párhuzamos szálak ütemezése

Ciklus kifejtés

Automatikus ciklusösszevonás