1 proPar Curso 14/15 4 2, 3, 2 3, 2 Computadores Paralelos Programación basada en paso de mensajes Técnicas básicas de programación paralela Agradable, Divide y Vencerás, Pipeline, Síncrona, Equilibrado de carga y Terminación Programación basada en memoria común Algoritmos y aplicaciones Ordenación, …

2 proPar Temario memComún-2Programación basada en memoria común Introducción El modelo de Pthread (POSIX) El modelo de OpenMP Pragmas Funciones Cilk++ CUDA El modelo de memoria transaccional

3 proPar Introducción memComún-3Procesador Memoria Modelo Hardware Modelo Software Sistema Operativo => Procesos | Threads Biblioteca junto con lenguaje secuencial => C++ con DAPPLE Lenguaje secuencial con directivas al compilador => OpenMP Ampliar sintaxis lenguaje secuencial con paralelismo => Cilk++ Lenguaje totalmente nuevo => Ada

4 proPar El modelo de Pthread (POSIX) memComún-4// Crea un thread que ejecuta la función indicada int pthread_create (pthread_t * thread, atributos, funcion, argumentos) // Devuelve el identificador del thread pthread_t pthread_self (void) // Termina la ejecución del thread int pthread_exit (void *estado) // El thread (padre) espera por la terminación de uno de sus hijos int pthread_join (pthread_t thread, void **estado) Primitivas básicas

5 proPar El modelo de Pthread (POSIX) memComún-5#include #include int main (int argc, char *argv[]){ pthread_t tA, tB; int datoA=0; datoB=0; pthread_create(&tA, NULL, A, NULL); pthread_create(&tB, NULL, B, NULL); pthread_join (tA, NULL); pthread_join (tB, NULL); exit (0); } void *B (void *basura) { datoB = 2000; sleep (5); printf (“datoA=%d\n”, datoA); pthread_exit (NULL); void *A (void *basura) { datoA = 1000; printf (“datoB=%d\n”, datoB); Ejemplo de uso

6 proPar El modelo de Pthread (POSIX) memComún-6Regiones críticas int pthread_mutex_lock (pthread_mutex_t *regionCritica) int pthread_mutex_unlock (pthread_mutex_t *regionCritica) Semáforos int sem_init (sem_t *S, int global, unsigned int valor) int sem_wait (sem_t *S) int sem_post (sem_t *S) int sem_destroy (sem_t *S)

7 proPar El modelo de Pthread (POSIX) memComún-7Práctica 5: piThreads.c static int sed1, sed2, dentro, veces, numThreads; static pthread_mutex_t RC; // Código del esclavo int main (int argc, char *argv[]) { pthread_t tids[4]; int i; struct timeval t0, t1, t; gettimeofday (&t0, NULL); veces = atoi(argv[1]); numThreads = atoi(argv[2]); sed1 = atoi(argv[3]); sed2 = atoi(argv[4]); dentro = 0; for (i=0; i pthread_create (&tids[i], NULL, esclavo, (void *) i); pthread_join (tids[i], NULL); gettimeofday (&t1, NULL); timersub(&t1, &t0, &t); printf ("Tiempo = %3ld:%3ld (seg:mseg)\n", t.tv_sec, t.tv_usec/1000); printf ("Valor estimado de pi: %7.5f\n", 4.0*dentro/veces); } %piThreads

8 proPar El modelo de Pthread (POSIX) memComún-8Práctica 5: piThreads.c void *esclavo(void *basura) { int i, dentroLocal, tid; unsigned short xi[3]; double x, y; tid = (int) basura; dentroLocal = 0; xi[0] = sed1; xi[1] = sed2; xi[2] = tid; for (i=tid; i x = erand48(xi); y = erand48(xi); if (x*x+y*y <= 1.0) dentroLocal++; } pthread_mutex_lock (&RC); dentro += dentroLocal; pthread_mutex_unlock (&RC); pthread_exit(NULL); exit (0);

9 proPar El modelo de OpenMP memComún-9Se basa en la idea de threads Thread maestro ¿ Cuántos threads ? Dual => 2 Tetra => 4 ? fork join % export OMP_NUM_THREADS=8 fork join Pragmas • Funciones OpenMP Directivas al compilador Estático Cuántos somos Quién soy Dinámico

10 proPar El modelo de OpenMP (Pragmas) memComún-10Paralelización de bucles 2 10 yo int i, V[N]; for (i=0; i < N; i++) V[i] = i; int i, V[N]; #pragma omp parallel for for (i=0; i < N; i++) V[i] = i; ? Reparto de iteraciones entre los threads activos int i, V[N]; for (i=1; i < N; i++) V[i] = V[i] + V[i-1]; ? schedule (static, 5) ¡ Iteraciones independientes !

11 proPar El modelo de OpenMP (Pragmas) memComún-11Paralelización de bucles: Variables comunes y locales “privadas” T0 T1 i i, V[N] int i, V[N]; #pragma omp parallel for for (i=0; i < N; i++) V[i] = i; local común Todo común salvo índice del for ¿Más variables locales? int f,c,M[F][C]; #pragma omp parallel for for (f=0; f < F; f++) for (c=0; c < C; c++) M[f][c] = f*C+c; private (c)

12 proPar El modelo de OpenMP (Pragmas) memComún-12Paralelización de bucles: Variables comunes y locales “privadas” for (i=0; i for (j=0; j #pragma omp parallel for for (j=0; j for (i=0; i for (j=0; j printf (“%4d”, V[i][j]); printf (“\n”); } {0 0 …… …… 1 1 …… 2 2 …… 2 2 …… 3 3 …… 3 3 ? { "@context": "http://schema.org", "@type": "ImageObject", "contentUrl": "http://slideplayer.es/4324995/14/images/12/proPar+El+modelo+de+OpenMP+%28Pragmas%29+memCom%C3%BAn-12.jpg", "name": "proPar El modelo de OpenMP (Pragmas) memComún-12", "description": "Paralelización de bucles: Variables comunes y locales privadas for (i=0; i

16 proPar El modelo de OpenMP (Funciones) memComún-16DINÁMICAS ESTÁTICAS // Número de procesadores de la máquina int omp_get_num_procs (void) // Número máximo de threads que se pueden activar int omp_get_max_threads (void) // OMP_NUM_THREADS // Activar un cierto número de threads <= máximo int omp_set_num_threads (int cuantos) // Número de threads activados int omp_get_num_threads (void) // Mi identificador => 0..omp_get_num_threads()-1 int omp_get_thread_num (void)

17 proPar El modelo de OpenMP (Funciones) memComún-17 void maestro(void) { printf (“Soy el maestro\n”); } void esclavo (int yo) { printf (“Soy el esclavo %d\n”, yo); numThreads = atoi(argv[1]); omp_set_num_threads (numThreads) #pragma omp parallel { int yo; yo = omp_get_thread_num(); if (yo == 0) maestro() else esclavo(yo);

18 proPar Cilk++ memComún-18Reto multicore  Solución comercial  Procedencia académica MIT  Leiserson Intel Jul/2009 Sep/2010 Cilk Plus

19 proPar Cilk++ memComún-19https://software.intel.com/en-us/intel-cilk-plus

20 proPar Cilk++ memComún-20https://www.cilkplus.org/build-gcc-cilkplus

21 proPar Cilk++ memComún-21

22 proPar Cilk++ memComún-22

23 proPar Cilk++ memComún-23

24 proPar Cilk++ memComún-24Facilidad de programación: Tres primitivas ¿ Qué hacer en paralelo ? ¡ Me olvido de la ineficiencia ! 5 4 3 2 1 1 2 3 int fib (int n) { if (n < 2) return (n); else { int x, y; x = cilk_spawn fib (n-1); y = fib (n-2); cilk_sync; return (x+y); } int fib (int n) { if (n < 2) return (n); else { int x, y; x = fib (n-1); y = fib (n-2); return (x+y); } Se indica qué no quién ¿Eficiente? Cilk++ Run Time System Planificador que reparte trabajo ¿Quién hace qué?

25 proPar Cilk++ memComún-25

26 proPar Cilk++ memComún-26Sin cambios en el código 1..16 Menos 2% pérdida en secuencial

27 proPar Cilk++ memComún-27Facilidad de programación: ¿Tres primitivas? int i, V[N]; for (i=0; i < N; i++) procesar (V[i]); int i, V[N]; cilk_for (i=0; i < N; i++) procesar (V[i]); Aplicado a la práctica 5: OpenMP y ordenar mediante cubetas 10.000 19.999 9.999 90.000 99.999 M-1 P0 P1 P9 Coger Ordenar Juntar

28 proPar Cilk++ memComún-28#include int cilk_main ( …. ) { ----- // Inicializar cilk_for (int j=0; j cogerYordenar (j); // Colocar cubetas en vDst ordenarCilk.c % ordenarCilk 4 -cilk_set_worker_count=4 cubetas 1 2 3 4 secuencial 96:685 48:352 32:291 24:202 Cilk++ 96:711 24:230 10:782 6:112 OpenMP 96:627 24:235 10:777 6:175

29 proPar CUDA (Arquitectura general) memComún-29Tarjetas Gráficas: Evolución (NVIDIA) 1999: GeForce 256 16 x 32 ..512 núcleos CUDA 2009: Fermi 2005: Play Station 3 15 x 192 núcleos CUDA 2012: Kepler 15 x 256 2006: CUDA 2012: #1 TOP500 2006 2008 2010 2012 2014 G80 GT200 Fermi Kepler Maxwell 128 256 512 1.536 2.048

30 proPar CUDA (Arquitectura general) memComún-30

31 proPar CUDA (Arquitectura general) memComún-31Tarjetas Gráficas: Comparativa FLOPS en CPU vs GPU

32 proPar CUDA (Arquitectura general) memComún-32Tarjetas Gráficas: Comparativa ancho de banda de memoria en CPU vs GPU

33 proPar CUDA (Arquitectura general) memComún-33Tarjetas Gráficas: En GPUs menos control, menos cache y más núcleos simples 4 cores a 2,66GHz 8MB de cache (L3) 6 GB de DRAM 731 millones transistores Core i7 3ª 2008 960 cores a 1,03GHz 384KB de cache (L2) 2 GB de DRAM 2.540 millones transistores GeForce GTX Threads Threads [Fractal]

34 proPar CUDA (Modelo de programación) memComún-34SIMT: La misma instrucción todos los threads ¿SIMD | PRAM | BSM? #pragma omp parallel for for (i=0; i < N; i++) C[i] = A[i] + B[i]; forall (i=0; i < N; i++) cilk_for(i=0; i < N; i++) 2 1 3 4 5 7 8 6 9 A B C + ¡ Ha desaparecido el for ! __global__ void sumaVector(float *Ad, float *Bd, float *Cd) { int i = threadIdx.x; Cd[i] = Ad[i] + Bd[i]; } sumaVector<<<1, N>>>(Ad,Bd,Cd); CPU GPU kernel Identificador único 0..N-1 Invocar kernel con N Threads

35 proPar CUDA (Modelo de programación) memComún-35Desde la CPU se pueden lanzar varios kernels: __global__ void sumaVector(float *Ad, float *Bd, float *Cd) { int i = threadIdx.x; Cd[i] = Ad[i] + Bd[i]; } sumaVector<<<1, N>>>(Ad,Bd,Cd); kernel Identificador único 0..N-1 Invocar kernel con N Threads CPU GPU Kernel_A CPU GPU Cd[i] = Ad[i] + B[i]; t0 t1 t2 t3 ti tN-1 cudaDeviceSynchronize(); Kernel_B Cd[i] = Ad[i] * B[i]; t0 t1 t2 t3 ti tM-1

36 proPar CUDA (Modelo de programación) memComún-36La GPU computa con datos que están en SU memoria => ¿Ad, Bd y Cd? __global__ void sumaVector(float *Ad, float *Bd, float *Cd) { int i = threadIdx.x; Cd[i] = Ad[i] + Bd[i]; } sumaVector<<<1, N>>>(Ad,Bd,Cd); kernel Identificador único 0..N-1 Invocar kernel con N Threads CPU GPU float *A, *B, *C; float *Ad, *Bd, *Cd; // Ubicar e inicializar A en CPU A = (float *) malloc (N * sizeof(float)); initVector (A, N, 0.001f); // Ubicar y transferir A a GPU cudaMalloc ((void**) &Ad, N); cudaMemcpy (Ad, A, N, cudaMemcpyHostToDevice); CPU GPU A A B Bd Ad Ad Cd cudaMemcpyDeviceToHost C  Cd C Cd C ¿CUDA 6?

37 proPar CUDA (Modelo de programación) memComún-37#define NBytes (N * sizeof(float)) __global__ void sumaVector(float *Ad, float *Bd,float *Cd) { int i = threadIdx.x; Cd[i] = Ad[i] + Bd[i]; } float *A, *B, *C; // Matrices a ubicar en CPU float *Ad, *Bd, *Cd; // Matrices a ubicar en GPU A = (float *) malloc (NBytes); // Ubicar A en CPU B = (float *) malloc (NBytes); // Ubicar B en CPU C = (float *) malloc (NBytes); // Ubicar C en CPU initVector (A, N, 0.001f); initVector (B, N, 0.002f); cudaMalloc ((void**) &Ad, NBytes); // Ubicar A en GPU cudaMalloc ((void**) &Bd, NBytes); // Ubicar B en GPU cudaMalloc ((void**) &Cd, NBytes); // Ubicar C en CPU cudaMemcpy (Ad, A, NBytes, cudaMemcpyHostToDevice); // A => GPU cudaMemcpy (Bd, B, NBytes, cudaMemcpyHostToDevice); // B => GPU sumaVector<<<1, N>>>(Ad, Bd, Cd); cudaDeviceSynchronize(); cudaMemcpy (C, Cd, NBytes, cudaMemcpyDeviceToHost); // Cd => CPU cudaFree (Ad); cudaFree(Bd); cudaFree(Cd);

38 proPar CUDA (Modelo de programación) memComún-38¿Tiempos de ejecución? microsegundos threads ¿Vectores más grandes? ? sumaVector<<<1, N>>>(Ad,Bd,Cd); numBloques numThreadsBloque ¡ Más bloques ! En nuestra GPU [3.0] el máximo es de 1024

39 proPar CUDA (Modelo de programación) memComún-39./sumVector 9600 #define anchoBloque 960 __global__ void sumaVector(float *Ad,float *Bd,float *Cd){ int i = threadIdx.x; Cd[i] = Ad[i] + Bd[i]; } sumaVector<<>>(Ad,Bd,Cd); blockIdx.x * anchoBloque + threadIdx.x; 10 960 ¡ Ojo ! 959 1 2 Bloque 0 Bloque 1 Bloque 9 threadIdx.x blockIdx.x Cd ----- 1 2 959 960 961 962 1919 8640 9599 ….

40 proPar CUDA (Modelo de programación) memComún-40¿Tiempos de ejecución “microsegundos”? CPU 8 96 785 5.709 53.731 GPU 42.152 42.728 42.958 45.901 71.375 1.000 10.000 ¡ Mucho tráfico de memoria entre CPU y GPU para poco cómputo ! 7.096 72.256 CPU => GPU CPU <= GPU GPU

41 proPar CUDA (Modelo de programación) memComún-41Las agrupaciones de bloques pueden ampliarse a 2D y 3D dim3 numBloques (3, 2); // 3 * 2 => 6 bloques dim3 threadsBloque (4, 3); // 4 * 3 => 12 threads mulMatriz<<>>(Ad,Bd,Cd); blockIdx.x blockIdx.y threadIdx.x threadIdx.y Block (2, 1) blockDim.x  3 blockDim.y  2

42 proPar CUDA (Modelo de programación) memComún-42Las agrupaciones de bloques pueden ampliarse a 2D y 3D Thread (3,0,0) (0,0,0) (0,1,0) (1,1,0) (2,1,0) (3,1,0) (1,0,0) (2,0,0) (0,0,1) (1,0,1) (2,0,1) (3,0,1) Grid Bloque (0, 0) (0, 1) (1, 0) (1, 1) (2, 0) (2, 1) 231-1 65535 Máximos [3.0] dim3 numBloques (3, 2, 0); dim3 threadsBloque (4, 2, 2); 1024 64 Máximos [3.0] maxThreadBloque  1024

43 proPar CUDA (Modelo de programación) memComún-43Multiplicación de matrices [cuadradas64]: C = A * B __global__ void mulMatriz ( float *Ad, float *Bd, float *Cd) { int fila = threadIdx.y; int col = threadIdx.x; int k; float valor = 0.0; for (k=0; k<64; k++) valor += Ad[fila, k] * Bd[k, col]; Cd[fila, col] = valor; } B A C fila 63 col dim3 thBlk(64, 64); mulMatriz <<<1, thBlk>>> (Ad, Bd, Cd); ¿ 64 * 64  4096 Threads ?

44 proPar CUDA (Modelo de programación) memComún-44Multiplicación de matrices [cuadradas64]: C = A * B __global__ void mulMatriz ( float *Ad, float *Bd, float *Cd) { int fila = blockIdx.y*16+threadIdx.y; int col = blockIdx.x*16+threadIdx.x; int k; float valor = 0.0; for (k=0; k<64; k++) valor += Ad[fila, k] * Bd[k, col]; Cd[fila, col] = valor; } B A C fila 63 col dim3 blks (4, 4) dim3 thBlk(16, 16); mulMatriz <<>> (Ad, Bd, Cd); ¿ 16 * 16  256 Threads ?

45 proPar CUDA (Modelo de programación) memComún-45#define N 64 // Dimension de las matrices cuadradas #define anchoBloque 16 // En 2D => 16x16 = 256 threads // void initMatriz (float *M, int card, float valor) { int i; for (i=0; i __global__ void mulMatrizKernel (float *Ad, float *Bd, float *Cd, int card) { int fil = blockIdx.y*anchoBloque + threadIdx.y; int col = blockIdx.x*anchoBloque + threadIdx.x; int k; float Cvalor = 0; for (k=0; k Cvalor += Ad[fil * card + k] * Bd[k * card + col]; Cd[fil * card + col] = Cvalor; }

46 proPar CUDA (Modelo de programación) memComún-46int main (int argc, char *argv[]){ struct timeval t0, tf, t; float *A, *B, *C; float *Ad, *Bd, *Cd; int sizeAByC, k; sizeAByC = N*N*sizeof(float); A = (float *) malloc (sizeAByC); B = (float *) malloc (sizeAByC); C = (float *) malloc (sizeAByC); initMatriz (A, N*N, 1.0f ); initMatriz (B, N*N, 0.01f); gettimeofday (&t0, NULL); // Transferir A y B a la GPU cudaMalloc ((void**) &Ad, sizeAByC); cudaMemcpy (Ad, A, sizeAByC, cudaMemcpyHostToDevice); cudaMalloc ((void**) &Bd, sizeAByC); cudaMemcpy (Bd, B, sizeAByC, cudaMemcpyHostToDevice);

47 proPar CUDA (Modelo de programación) memComún-47// Ubicar C en la CPU cudaMalloc ((void**) &Cd, sizeAByC); // Invocar al kernel dim3 dimGrid (N/anchoBloque, N/anchoBloque); dim3 dimBlock(anchoBloque, anchoBloque); mulMatrizKernel<<>>(Ad, Bd, Cd, N); cudaDeviceSynchronize(); // Transferir C desde la GPU cudaMemcpy (C, Cd, sizeAByC, cudaMemcpyDeviceToHost); // Liberar matrices en la GPU cudaFree(Ad); cudaFree(Bd); cudaFree(Cd); gettimeofday (&tf, NULL); timersub (&tf, &t0, &t); printf ("Tiempo = %ld:%ld \n", t.tv_sec, t.tv_usec); }

48 proPar CUDA (Modelo de programación) memComún-48¿Tiempos de ejecución “seg:miliseg” ? CPU 0:022 0:183 1:777 14:549 GPU 0:043 0:045 0:057 0:145 160 x 160 320 x 320 640 x 640 1.280 x 1.280

49 proPar CUDA (Modelo de programación) memComún-49Representación del conjunto de Mandelbrot 960 x 960 => pixels colores mandelsecBis 30:892 mandelpar –np 4 8:078 CPU *25,7 *6,7 mandelparBis 1:202 GPU *7,4 ¡ Todo negro !

50 proPar CUDA (Modelo de programación) memComún-50#define FILAS #define COLUMNAS 960 #define anchoBloque 16 // En 2D => 16x16 = 256 threads static int *colores, *coloresd; // Matrices de colores static int sizeColores; // Programa principal sizeColores = FILAS * COLUMNAS * sizeof(int); colores = (int *) malloc (sizeColores); // Ubico matriz de colores en GPU cudaMalloc ((void**) &coloresd, sizeColores); // Invocar al kernel dim3 dimGrid (FILAS/anchoBloque, COLUMNAS/anchoBloque); dim3 dimBlock(anchoBloque, anchoBloque); mandelKernel<<>>( planoFactorX, planoFactorY, planoVx, planoVy, mapiNumColoresDefinidos(), coloresd); 60 * 60 => bloques de 256 Threads ¡ Threads !

51 proPar CUDA (Modelo de programación) memComún-51__global__ void mandelKernel(…){ int fila, columna, i; double X, Y; double pReal, pImag = 0.0; // Parte real X e imaginaria Y double pRealAnt, pImagAnt, distancia; // Determinar pixel fila = blockIdx.y*anchoBloque + threadIdx.y; columna = blockIdx.x*anchoBloque + threadIdx.x; // planoPixelAPunto X = (pFactorXd * (double) columna) + pVxd; Y = (pFactorYd * ((double) (FILAS-1) - (double) fila)) + pVyd; // Se evalua la formula de Mandelbrot i=0; do { pRealAnt = pReal; pImagAnt = pImag; pReal = ((pRealAnt*pRealAnt) - (pImagAnt*pImagAnt)) + X; pImag = (2.0 * (pRealAnt*pImagAnt)) + Y; distancia = pReal*pReal + pImag*pImag; } while ((++i < maxIteracionesd) && (distancia <= 4.0)); if (i == maxIteracionesd) i = 0; coloresd[fila * COLUMNAS + columna] = i; }

52 proPar CUDA (Arquitectura detalle) memComún-52384KB

53 proPar CUDA (Arquitectura detalle) memComún-53CPU Grid3 + Bloques3 Warps 32 Threads Todos arrancan en I0 del kernel [síncronos] If_Then_Else => Se serializan ramas: Threads activos e inactivos en cada rama Al final del kernel barrera entre ellos Cada SMX ejecuta simultáneamente 6 warps Cada SMX puede albergar 64 warps En cada ciclo se eligen los 6 warps a ejecutar 2.048 Threads Registros

54 proPar El modelo de memoria transaccional memComún-54Memoria común Más de 50 años de programación concurrente y … ¿Regiones críticas: locks, condiciones, monitores, …? void mover (lOrg, lDst, clave){ LOCK(); tmp = lOrg.sacar(clave); lDst.meter(tmp); UNLOCK(); } void mover (lOrg, lDst, clave){ LOCK(lOrg); LOCK(lDst); tmp = lOrg.sacar(clave); lDst.meter(tmp); UNLOCK(lDst); UNLOCK(lOrg); } Grano grueso  Serialización ¡ Posible deadlock ! mover (a, b, clave1);  Thread.i mover (b, a, clave2);  Thread.j ¿Cómo componer métodos con locks internos?

55 proPar El modelo de memoria transaccional memComún-55Los multicore ponen aún más en evidencia el problema de los locks ¿Cut & Paste?: Bases de Datos y transacciones [ACID] void mover (lOrg, lDst, clave){ atomic { tmp = lOrg.sacar(clave); lDst.meter(tmp); } Qué Cómo ¡Más simple! Atomicidad: Todo o nada Commit (OK) o abort (retry) Conjuntos de Read/Write Conflictos Aislamiento: Valores intermedios invisibles Serialización ¿Eficiente?

56 proPar El modelo de memoria transaccional memComún-56¡ No todo está resuelto ! bool flagA = flagB = 0; atomic { while (flagA == 0); flagB=1; } Thread.i flagA=1; while (flagB == 0); Thread.j Se han propuesto más primitivas: retry, watch, orElse, … Adaptable al grano fino: int i, vectorA[101], vectorB[1000]; for (i=0; i<1000; i++) atomic { vectorA[vectorB[i]]++; }

57 proPar El modelo de memoria transaccional memComún-57Algunas cuestiones de diseño ¿Atomicidad débil o fuerte? ¡ Sólo en el código transaccional ! ¿Anidamiento de transacciones? Tex Tin Planas Tex Tin Cerradas Tex Tin Abiertas abort control control abort abort abort commit commit commit visible visible visible

58 proPar El modelo de memoria transaccional memComún-58

59 proPar El modelo de memoria transaccional memComún-59Una implementación STM 4.0 “STM: Why is it only a research toy?”. Cascaval, … ACM Sep/2008

60 proPar El modelo de memoria transaccional memComún-60Una implementación HyTM: SUN Rock 16 cores Soporte HTM ¿Sep/Nov 2009? “Early Experience with a Commercial HTM Implementation” D. Dice, … ASPLOS’09, 7-11 de Marzo 2009

61 proPar El modelo de memoria transaccional memComún-61Una implementación HTM: IBM Blue Gene/Q Compute Chip 18 cores PowerPC A2 L1 (16+16K) privada L2 2MB (multiversión) Instalado en el Sequoia #1 => #2 (Nov/2012)

62 proPar El modelo de memoria transaccional memComún-62Una implementación HTM: Intel Haswell Jun/2013 => TSX

63 proPar El modelo de memoria transaccional memComún-63Una implementación HTM: Intel Haswell Jun/2013 => TSX FIN