-
Notifications
You must be signed in to change notification settings - Fork 0
Expand file tree
/
Copy pathindex.html
More file actions
1627 lines (1481 loc) · 71.5 KB
/
Copy pathindex.html
File metadata and controls
1627 lines (1481 loc) · 71.5 KB
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136
137
138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
182
183
184
185
186
187
188
189
190
191
192
193
194
195
196
197
198
199
200
201
202
203
204
205
206
207
208
209
210
211
212
213
214
215
216
217
218
219
220
221
222
223
224
225
226
227
228
229
230
231
232
233
234
235
236
237
238
239
240
241
242
243
244
245
246
247
248
249
250
251
252
253
254
255
256
257
258
259
260
261
262
263
264
265
266
267
268
269
270
271
272
273
274
275
276
277
278
279
280
281
282
283
284
285
286
287
288
289
290
291
292
293
294
295
296
297
298
299
300
301
302
303
304
305
306
307
308
309
310
311
312
313
314
315
316
317
318
319
320
321
322
323
324
325
326
327
328
329
330
331
332
333
334
335
336
337
338
339
340
341
342
343
344
345
346
347
348
349
350
351
352
353
354
355
356
357
358
359
360
361
362
363
364
365
366
367
368
369
370
371
372
373
374
375
376
377
378
379
380
381
382
383
384
385
386
387
388
389
390
391
392
393
394
395
396
397
398
399
400
401
402
403
404
405
406
407
408
409
410
411
412
413
414
415
416
417
418
419
420
421
422
423
424
425
426
427
428
429
430
431
432
433
434
435
436
437
438
439
440
441
442
443
444
445
446
447
448
449
450
451
452
453
454
455
456
457
458
459
460
461
462
463
464
465
466
467
468
469
470
471
472
473
474
475
476
477
478
479
480
481
482
483
484
485
486
487
488
489
490
491
492
493
494
495
496
497
498
499
500
501
502
503
504
505
506
507
508
509
510
511
512
513
514
515
516
517
518
519
520
521
522
523
524
525
526
527
528
529
530
531
532
533
534
535
536
537
538
539
540
541
542
543
544
545
546
547
548
549
550
551
552
553
554
555
556
557
558
559
560
561
562
563
564
565
566
567
568
569
570
571
572
573
574
575
576
577
578
579
580
581
582
583
584
585
586
587
588
589
590
591
592
593
594
595
596
597
598
599
600
601
602
603
604
605
606
607
608
609
610
611
612
613
614
615
616
617
618
619
620
621
622
623
624
625
626
627
628
629
630
631
632
633
634
635
636
637
638
639
640
641
642
643
644
645
646
647
648
649
650
651
652
653
654
655
656
657
658
659
660
661
662
663
664
665
666
667
668
669
670
671
672
673
674
675
676
677
678
679
680
681
682
683
684
685
686
687
688
689
690
691
692
693
694
695
696
697
698
699
700
701
702
703
704
705
706
707
708
709
710
711
712
713
714
715
716
717
718
719
720
721
722
723
724
725
726
727
728
729
730
731
732
733
734
735
736
737
738
739
740
741
742
743
744
745
746
747
748
749
750
751
752
753
754
755
756
757
758
759
760
761
762
763
764
765
766
767
768
769
770
771
772
773
774
775
776
777
778
779
780
781
782
783
784
785
786
787
788
789
790
791
792
793
794
795
796
797
798
799
800
801
802
803
804
805
806
807
808
809
810
811
812
813
814
815
816
817
818
819
820
821
822
823
824
825
826
827
828
829
830
831
832
833
834
835
836
837
838
839
840
841
842
843
844
845
846
847
848
849
850
851
852
853
854
855
856
857
858
859
860
861
862
863
864
865
866
867
868
869
870
871
872
873
874
875
876
877
878
879
880
881
882
883
884
885
886
887
888
889
890
891
892
893
894
895
896
897
898
899
900
901
902
903
904
905
906
907
908
909
910
911
912
913
914
915
916
917
918
919
920
921
922
923
924
925
926
927
928
929
930
931
932
933
934
935
936
937
938
939
940
941
942
943
944
945
946
947
948
949
950
951
952
953
954
955
956
957
958
959
960
961
962
963
964
965
966
967
968
969
970
971
972
973
974
975
976
977
978
979
980
981
982
983
984
985
986
987
988
989
990
991
992
993
994
995
996
997
998
999
1000
<!doctype html>
<html lang="en">
<head>
<meta charset="utf-8">
<title>Introduction to GPU programming</title>
<meta name="description" content="A framework for easily creating beautiful presentations using HTML">
<meta name="author" content="Hakim El Hattab">
<meta name="apple-mobile-web-app-capable" content="yes">
<meta name="apple-mobile-web-app-status-bar-style" content="black-translucent">
<meta name="viewport"
content="width=device-width, initial-scale=1.0, maximum-scale=1.0, user-scalable=no, minimal-ui">
<link rel="stylesheet" href="css/reveal.css">
<link rel="stylesheet" href="css/theme/black.css" id="theme">
<!-- Code syntax highlighting -->
<link rel="stylesheet" href="lib/css/zenburn.css">
<!-- Printing and PDF exports -->
<script>
var link = document.createElement('link');
link.rel = 'stylesheet';
link.type = 'text/css';
link.href = window.location.search.match(/print-pdf/gi) ? 'css/print/pdf.css' : 'css/print/paper.css';
document.getElementsByTagName('head')[0].appendChild(link);
</script>
<!--[if lt IE 9]>
<script src="lib/js/html5shiv.js"></script>
<![endif]-->
</head>
<body>
<div class="reveal">
<!-- Any section element inside of this container is displayed as a slide -->
<div class="slides">
<section>
<h1>Introduction to GPU programming</h1>
<h3>29.08.2013</h3>
<p>
<small>Eirik Ola Aksnes</small>
</p>
</section>
<section>
<h2>Agenda</h2>
<ul>
<li>What are GPUs?</li>
<li>Where can GPUs be found?</li>
<li>Historical motivation - the video game industry</li>
<li>GPGPU - general-purpose computing on GPUs</li>
<li>CPU vs. GPU</li>
<li>Data parallelism</li>
<li>Nvidia CUDA</li>
<li>CUDA example - Vector addition</li>
<li>Final comments</li>
</ul>
</section>
<section>
<h2>What are GPUs?</h2>
<ul>
<li>GPUs are highly parallel, multithreaded, many-core processors</li>
<li>Hundreds of cores</li>
<li>Thousands of concurrent threads</li>
<li>First GPU - Nvidia's GeForce 256 (1999)</li>
<li>Vendors - NVIDIA, AMD, Intel</li>
</ul>
<aside class="notes">
<ul>
<li>En GPU er en svært parallell, flertrådet, mange-kjerner prosessorer.</li>
<li>Den består av hundrevis av kjerner, som kan kjøre tusenvis av tråder samtidig.</li>
<li>Det var NVIDIA som startet å bruke utrykket GPU...</li>
</ul>
</aside>
</section>
<section>
<h2>Where can GPUs be found?</h2>
<ul>
<li>Mobile phones</li>
<li>Personal computers</li>
<li>Clusters</li>
<li>Supercomputers</li>
<li>Game consoles</li>
</ul>
</section>
<section>
<h2>Historical motivation - the video game industry</h2>
<blockquote>
Constantly pushes to improve the ability to perform massive numbers of floating point calculations in
video games (multi-billion dollar industry!)
</blockquote>
<aside class="notes">
Historisk sett så er utviklingen av GPUer drevet av spill industrien.
<ul>
<li>En GPU er opprinnelig laget for å effektivt utføre beregninger som kreves av 3D grafikk.</li>
<li>Det er en prossesor som er spesielt designet for beregnings intensive problemer.</li>
</ul>
</aside>
</section>
<section>
<h2>1981 - 3D Monster Maze</h2>
<img data-src="images/3d-monster-mace.png"/>
<aside class="notes">
3D monster maze er verdens første 3D-spill for en hjemme-PC.
</aside>
</section>
<section>
<h2>2013 - Tom Clancy’s The Division</h2>
<img data-src="images/tom-clancys-the-division.png"/>
<aside class="notes">
Tom Clancy’s fikk beste pris for 3D grafikk i 2013. Som vi ser, så har det skjedd mye siden 1981.
</aside>
</section>
<section>
<h2>GPGPU - General-purpose computing on GPUs</h2>
<ul>
<li>GPUs can be used for more than just graphic rendering</li>
<li>GPUs can accelerate applications in a variety of disciplines
<ul><li>Big speedup compared to CPUs in some cases</li></ul>
</li>
<li>In the beginning graphic APIs, like OpenGL, was used to do GPGPU (big hack)
<ul>
<li>Difficult to develop, debug, and optimize</li>
</ul>
</li>
<li>Improvements in hardware and software has made GPGPU easier
<ul>
<li>Programmability</li>
</ul>
</li>
</ul>
<aside class="notes">
<ul>
<li>..de kan brukes til å utføre beregninger tradisjonelt håndtert av CPUer, til å utføre generelle
beregninger.
</li>
<li>Det å gjøre GPU programmering med OpenGL er vanskelig: Blant annet så krever det kunnskap om
grafikk APIer, noe som er unødvendig for å programmere ikke grafiske applikasjoner.
</li>
<li>Det har blitt lettetere å programmere dem...</li>
</ul>
</aside>
</section>
<section>
<h2>History of GPGPU computing</h2>
<img data-src="images/history-of-gpgpu-programming.png"/>
<aside class="notes">
Hvis vi ser på historien til GPGPU programmering:
<br />
<br />
Vi set at i begynnelsen ble grafikk APIer brukt:
<ul>
<li>OpenGL</li>
<li>DirectX (Microsoft)</li>
</ul>
<br/>
Rundt ~ 2005 - Ulike abstraksjoner - Høy nivå (tredjeparts) språk som abstrahert bort grafikken.
<ul>
<li>BrookGPU - Et generelt rammeverk for GPU programmering utviklet av Stanford University.
Programmer skrives i språket Brook som er en utvidet versjon av standard C.
</li>
</ul>
<br/>
Rundt ~ 2006 - Ble det sluppet fra leverandører, dedikerte språk for GPGPU programmering.
<ul>
<li>I dag skal vi se på NVIDIA sitt som heter CUDA</li>
<li>AMD - Close-to-Metal (CTM)</li>
<li>DirectCompute - Microsoft</li>
<li>OpenCL bør også nevnes:</li>
<ul>
<li>OpenCL står for Open Computing Language</li>
<li>Tilsvarende CUDA, men er en åpen standard</li>
<li>Opprinnelig utviklet av Apple, men er nå vedlikeholdt av Khronos gruppen (same gruppe som
vedlikeholder OpenGL)
</li>
<li>En stor fordel med OpenCL er at man ikke bare er begrenset til en GPU leverandør, man kan
faktisk også bruke CPUer, og potensielt en rekke andre enheter. For å gjøre data-parallelle
beregninger.....
</li>
</ul>
</ul>
</aside>
</section>
<section>
<section>
<h1>CPU vs. GPU</h1>
</section>
<section>
<h2>Floating-point performance</h2>
<img height="500" data-src="images/cpu-vs-gpu-floating-point-performance.png"></img>
<aside class="notes">
Hvorfor har man startet å bruke GPUer til andre ting en grafikk?
<ul>
<li>En av grunnene er at de har svært god ytelse i forhold til CPUer.</li>
</ul>
Her ser vi en graf som viser hva NVIDIA sine GPUer og Intel sine CPUer klarer av antall
flyttallsoperasjoner per sekund. Det vi ser ut ifra grafen er at:
<ul>
<li>GPUene er først å fremst raskere en CPuene, men vi kan også se at</li>
<li>ytelsen til GPUer øker raskere enn ytelsen til CPUer.</li>
<li>De har en større stigende kurve...</li>
<li>Det er viktig å huske at dette er den teoretiske beste ytelsen, og at det er skal veldig
mye til for å
få til en så god ytelsen…
</li>
</ul>
</aside>
</section>
<section>
<h2>Memory bandwidth</h2>
<img height="500" data-src="images/cpu-vs-gpu-memory-bandwidth.png"/>
<aside class="notes">
<ul>
<li>En annen fordel er at GPUer har høy minnebåndbredde...</li>
<li>Ofte er applikasjoner bundet av minne, så en høy minnebåndbredde er viktig for å oppnå høy
ytelse for mange problemer....
</li>
</ul>
</aside>
</section>
<section>
<h2>Design</h2>
<table>
<thead>
<tr>
<th>CPU</th>
<th>GPU</th>
</tr>
</thead>
<tbody>
<tr>
<td>Few complex and speedy cores</td>
<td>Many simple and slow cores</td>
</tr>
<tr>
<td>Each core has own sophisticated control logic (independent execution)</td>
<td>Groups of compute cores share control logic</td>
</tr>
<tr>
<td>Memory latency hidden by cache and prefetching</td>
<td>Memory latency hidden by swapping threads. Massive multithreading. Very fast context
switching
</td>
</tr>
</tbody>
</table>
<aside class="notes">
Tenkte å gå igjennom en del punkter som skiller CPUer og GPUer i design…
<ul>
<li>En CPU har noen få kjerner (typisk to, fire eller åtte kjerner) som er veldig komplekse og
raske.
</li>
<li>En GPU har mange (hundrevis) enkle og trege kjerner.</li>
</ul>
<ul>
<li>Hver CPU kjerne har egen avansert kontroll logikk.</li>
<li>Hos en GPU så deler grupper av kjerner kontrol logikk.</li>
</ul>
<ul>
<li>En CPU skjuler minne ventetid ved bruk av “cacheing” og “prefetching”.</li>
<li>En GPU skjuler minne ventetid ved å bytte mellom tråder som kjører.</li>
<li>Hvis en tråd venter på minne aksess, så bytter GPUen til annen tråd, som kan gjøre
beregninger
istedenfor.
</li>
<li>Det å bytte mellom tråder går veldig fort (man har veldig rask kontekst svitsjing).</li>
</ul>
<!--Backup:
Fordi den samme beregningen utføres for hvert dataelement, er det en lavere krav til kontrol logikk for en GPU.
High performance on a single thread of execution
En CPU er designet for å minimere kjøretiden til en enkel tråd.
En GPU er designet for å få prossesert ferdig flest mulig tråder innenfor ett gitt tidpunkt.
High throughput of parallel threads of execution-->
</aside>
</section>
<section>
<h2>Transistors</h2>
<!--<p>The GPU devotes more transistors to data processing. Processing units are shown in green.</p>-->
<img height="200" data-src="images/cpu-vs-gpu-transistors.png"/>
<ul>
<li>
CPU: Uses large fraction of the chip area for control logic and cache
</li>
<li>
GPU: Uses most of the chip area for data processing units
</li>
</ul>
<aside class="notes">
<ul>
<li>Oppsumert, så er GPUen utformet slik at flere transistorer er viet til å gjøre beregninger
enn
til data caching og flytkontroll.
</li>
<li>De grønne boksene er prosesserings enheter, mens de gule og oransje boksene er data caching
og
flytkontroll.
</li>
<li>En CPU er designet for å kjøre noen få tråder veldig raskt.</li>
<li>En GPU er designet for å kjøre tusenvis av tregere tråder samtidig.</li>
<!--Extra:
For the CPU a lot of transistor space is dedicated to complex instruction
level parallelism such as instruction pipelining, branch prediction,
speculative execution, and out-of-order execution…-->
</ul>
</aside>
</section>
<section>
<h2>Latency vs. Throughput</h2>
<p>What is better? What do you need?</p>
<img height="200" data-src="images/latency-vs-throughput.png"/>
<table>
<tbody>
<tr>
<td>CPU - Low latency</td>
<td>GPU - High Throughput</td>
</tr>
<tr>
<td>E.g. deliver a package as soon as possible</td>
<td>E.g. deliver many packages within a reasonable time</td>
</tr>
</tbody>
</table>
<aside class="notes">
<ul>
<li>En CPU kan sammenlignes med to sportsbiler.</li>
<li>En GPU kan sammenlignes med mange scootere.</li>
</ul>
</li>
Hva er bedre? Det er avhengig av hva du trenger?
<ul>
<li>En CPU er optimalisert for lav ventetid, altså designet for å minimere kjøretiden til en
enkel
tråd.
</li>
<li>En GPU er optimalisert for høy gjennomstrømning, altså designet for å få prossesert ferdig
flest
mulig tråder innenfor en gitt tid.
</li>
</ul>
<ul>
<li>Beregn en jobb så fort som mulig -> CPU, kan sammenlignes med f.eks levere en pakke så snart
som
mulig.
</li>
<li>Beregn mange jobber innenfor en gitt tid -> GPU, kan sammenlignes med f.eks levere mange
pakker
innenfor en gitt tid.
</li>
</ul>
<ul>
<li>Det er mange applikasjoner hvor det å optimalisere for høy gjennomstrømning er mest
viktig.
</li>
<li>I f.eks bilde prosessering er man mer opptatt av antall piksler per sekund, enn ventetiden
av en
hvilken som helst piksel.
</li>
<li>Man er villig til å la tiden det tar å prosessere en piksel ta dobbel så lang tid, så lenge
vi
får prosessert flere pikseler på innenfor en git tid.
</li>
<li>For dette vil en GPU være gunstig.</li>
</ul>
Latency:
<ul>
<li>Hvor lang tid tar det å bli ferdig med en oppgave</li>
<li>Time (Seconds)</li>
</ul>
Throughput:
<ul>
<li>Antall oppgaver ferdig i en gitt tid</li>
<li>Stuff / Time (Jobs/Hours)</li>
<li>Image processing - Pixels / Seconds</li>
</ul>
<!--Backup:
High throughput and reasonable latency: Compute many jobs within a reasonable timeframe. E.g deliver
many packages within a reasonable timescale.
Low latency and reasonable throughput: Compute a job as fast as possible. E.g deliver a package as soon
as possible.
Pizza example:
http://on-demand.gputechconf.com/siggraph/2013/presentation/SG3112-GRID-State-of-the-Art-Virtualized-Graphics.pdf-->
</aside>
</section>
</section>
<section>
<h2>Data parallelism</h2>
<img height="300" data-src="images/data-parallelism.png"/>
<ul>
<li>Multiple processing units performs in parallel the same operation on different data elements</li>
<li>Flynn's taxonomy: Single-Instruction Multiple-Data (SIMD)</li>
<li>NVIDIA: Single-Instruction Multiple-Threads (SIMT)</li>
</ul>
<aside class="notes">
<ul>
<li>En GPU er spesielt godt egnet til å løse problemer som kan uttrykkes som data-parallelle
beregninger.
</li>
<li>I 3D grafikk er det veldig mange beregninger som utføres for hvert sekund, men i tillegg til
det, så er det ofte den samme beregningen som utføres, bare på forskjellige data elementer.
F.eks det å fargelege alle objekter i en 3D verden.
</li>
<li>
Data paralilittet vil si at man har mange prosesserings enheter…..
</li>
<li>NVIDIA kaller sin data-parallelle programmings model for SIMT.</li>
</ul>
<!--Backup:
It can be said that SIMT is a more flexible SIMD.
SIMT er en mer fleksibel enn SIMD, den støtter blant annet forgrening (f.eks. if-test, for-løkke
el.l.)...
SIMT and SIMD both approach parallelism through broadcasting the same instruction to multiple execution
units. This way, you replicate the execution units, but they all share the same fetch/decode hardware.
Indeed, although typically every thread will run identical functions, the functions themselves can
condition on thread identifiers and data so that different instructions are executed in some threads.
However, in SIMD architectures this leads to a performance hit since computation only occurs in parallel
when the same instructions are being performed
More specifically, the GPU is especially well-suited to address problems that can be expressed as
data-parallel computations.
En GPU er altså laget for å utføre veldig mange identiske beregninger på forskjellige data elementer
veldig raskt.
URLs:
http://www.yosefk.com/blog/simd-simt-smt-parallelism-in-nvidia-gpus.html-->
</aside>
</section>
<section>
<h2>Data parallelism - Image example</h2>
<img height="400" data-src="images/data-parallelism-threads.png"/>
<aside class="notes">
<ul>
<li>Hvis vi ser på ett enkelt eksempel…</li>
<li>...prosessering av piksler i ett bilde..</li>
<li>...vi antar at vi kan prosessere piksler uavhengig av hverandre...</li>
<li>Hvis vi ser på CPU eksempelet først, så er det typisk en tråd som går igjennom alle pikslene, en
etter en, via en for-løkke. F.eks for å konvertere piksler fra farge verdier til svart hvit…
</li>
<li>Hvis vi ser på GPU eksempelet, så er ideen at man har mange tråder som prosesserer piksel i
parallell.
</li>
<li>Hver tråd vil utføre den samme operasjonen, bare på forskjellige data elementer...</li>
<li>Hvor mange tråder som vil bli prosessert i parallell er avhengig av hvor mange kjerner GPUen man
benytter har.
</li>
</ul>
<!--Functions that are executed many times, but independently on different data, are prime candidates:
I.e. body of for-loops
A lot of algorithms perform computations by iteratively traversing a large data set. This is often
handled by a loop. In a sequential program this is executed in a sequential manner. If there are no
data dependencies, this loop could be performed in any arbitrary order and would be well suited for
parallel computing.-->
</ul>
</aside>
</section>
<section>
<section>
<img data-src="images/nvidia-cuda.jpg" />
<aside class="notes">
<ul>
<li> Da skal vi begynne å se på hvordan vi kan programmere NVIDIA sine GPUer...</li>
<li>Det blir en del teori først nå, før vi skal ta for oss ett kode eksempel (hvor vi skal legge
sammen to vektorer)...
</li>
</ul>
</aside>
</section>
<section>
<h2>What is CUDA?</h2>
<ul>
<li>Compute Unified Device Architecture</li>
<li> Nvidia introduced CUDA in 2006</li>
<li>Specially designed for GPGPU</li>
<li>Only supported by Nvidia graphics cards</li>
<li>Write code in C/C++, Java, Python, Fortran, Perl...</li>
<li>You do not need parallel programming experience</li>
<li>Requires no knowledge of graphics APIs</li>
<li>Access to native instructions and memory</li>
</ul>
</section>
<section>
<h2>Get started - CUDA Toolkit</h2>
<ul>
<li>Driver</li>
<li>Compiler nvcc</li>
<li>Development, profiling and debugging tools</li>
<li>Various libraries</li>
<li>Programming guides, and API reference</li>
<li>Example codes</li>
</ul>
<aside class="notes">
<ul>
<li>Før man kan starte å programmere med CUDA, så må man laste ned “CUDA Toolkit”...</li>
<li>Inneholder en del biblioteker som kan vurdere å bruke....</li>
<li>Eksempel koden er veldig fint å se på....</li>
</ul>
</aside>
</section>
<section>
<h2>Get started</h2>
<img data-src="images/get-started.png"/>
<aside class="notes">
Det er tre måter man kan starte å bruke CUDA:
<ul>
1) På venstre her har vi det å bruke biblioteker:
<ul>
<li> cuFFT, - Fast Fourier Transforms Library</li>
<li>cuBLAS - Complete BLAS bibliotek</li>
</ul>
2) I midten har vi OpenACC som er svært lik OpenMP programmering, for de som har gjort det.
<ul>
<li>Man legger til PRAGMAer (direktiver) i koden, som forteller kompilatoren at den skal
parallisere den kommenterte delen av koden for deg.
</li>
<li>F.eks hvis man skriver #pragma acc parallel foran en for-løkker, skal kompliatoren
automatisk paralllisere for-løkken for deg...
</li>
</ul>
3) Men i dag skal vi skal bruke programmeringsspråk, som er den mest kraftfulle og fleksible
måten.
Vi vil bruke C som språk...
</ul>
</aside>
</section>
<section>
<h2>CUDA C Programming Guide</h2>
<p>The advent of multicore CPUs and many-core GPUs means that mainstream processor chips are now
parallel
systems. Furthermore, their parallelism continues to scale with Moore’s law. The challenge is to
develop
application software that <u>transparently</u> scales its parallelism to leverage the increasing
number
of
processor cores...</p>
<p>At its core are three key abstractions – <u>a hierarchy of thread groups, shared memories, and
barrier
synchronization...</u></p>
<p>...<u>data parallelism</u>...</p>
<aside class="notes">
CUDA C programmerings veiledning...
<br />
The advent of: Ankomsten av flerkjernet CPUer og mange kjernet
GPUer.
</aside>
</section>
<section>
<h2>CUDA terminology</h2>
<img data-src="images/cuda-terminology.png"/>
<aside class="notes">
Begynner med litt terminologi...
</aside>
</section>
<section>
<h2>Heterogeneous programming</h2>
<div class='left' style='float:left;width:48%'>
<ul><li>Serial code → Host</li> <li>Parallel code → Device</li></ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="550" data-src="images/cuda.png"/>
</div>
<aside class="notes">
<ul>
<li>Det vi snakker om i dag er GPU programmering, også kalt heterogen programmering.</li>
<li>Som er koordinering av to eller flere forskjellige prosessorer, av forskjellige arkitektur
typer, for å utføre en felles oppgave.
</li>
<li>Man kombinerer det beste fra to verdener: CPU + GPU</li>
<li>CPU og GPU er en kraftfull kombinasjon fordi CPUer er optimalisert for seriell prosessering,
mens GPU består er optimalisert for parallell prosessering.
</li>
<li>Parallelle deler av et program kjøres på GPUen som “kernels”.</li>
</ul>
<!--Backup:
Heterogeneous computing is the coordination of 2 or more different processors, of different architecture types, to perform a computational task.
Heterogeneous programming = GPU Programming
CPU + GPU is a powerful combination because CPUs consist of a few cores optimized for serial processing, while GPUs consist of thousands of smaller, more efficient cores designed for parallel performance. Serial portions of the code run on the CPU while parallel portions run on the GPU.
Combining the best out of two world: CPU + GPU
Parallel portions of an application are executed on the device as kernels.
In addition, the combination of CPU and GPU heterogeneous computing framework also will be able to further improve efficiency and reduce power-->
</aside>
</section>
<section>
<h2>Kernels</h2>
<p>C functions, that when called, are executed N times in parallel by N different threads on the device.
As
opposed to only once, like a regular C functions.</p>
<img height="400" data-src="images/cuda-kernels.png"/>
<aside class="notes">
Hva er en kernel?
<ul>
<li>det er en C-funksjon, som, når den blir kalt, blir utført N ganger i parallell av N
forskjellige
tråder, i motsetning til en vanlig C-funksjon, som bare blir utført en gang.
</li>
<li>...så alt som skjer inne funksjonen, blir utført N antall ganger...</li>
<li>__global__ definerer at funksjonen er en kernel, og at funksjonen skal bli kjørt på GPUen...
</li>
<li>Antallet ganger “kernelen” blir utført (eller antall tråder), er spesifisert med en spesiell
syntaks… <<<>>>...vi skal komme tilbake til det snart...
</li>
</ul>
</aside>
</section>
<section>
<h2>Thread hierarchy</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>A kernel is executed by thousands of threads in parallel, organized as a hierarchy.</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="500" data-src="images/cuda-thread-hieachy.png"/>
</div>
<aside class="notes">
<ul>
<li>En “kernel” blir utført av tusenvis av tråder i parallell..</li>
<li>...i form av ett hierarki!</li>
</ul>
</aside>
</section>
<section>
<h2>Thread hierarchy - Blocks</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Threads are grouped into blocks (1D, 2D or 3D)
<ul>
<li>Each thread has its own local block ID</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="500" data-src="images/cuda-thread-hieachy.png"/>
</div>
<aside class="notes">
<ul>
<li>En “kernel” blir utført av tusenvis av tråder i parallell..</li>
<li>...i form av ett hierarki!</li>
<li>Som vi ser på bildet...</li>
<li>Tråder er gruppert i blokker (1D, 2D, eller 3D)</li>
<li>Hver tråd har en unik lokal ID i blokken sin...</li>
</ul>
<!--Backup:
Max number of threads per block ~ 512 or 1024
Every thread operates on a specific data element
URLs:
http://www.pgroup.com/lit/articles/insider/v2n1a5.htm-->
</ul>
</aside>
</section>
<section>
<h2>Thread hierarchy - Grids</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>
Blocks are grouped into a grid (1D, 2D or 3D)
<ul>
<li>Each block as its own ID</li>
<li>All blocks in a grid have the same dimension</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="500" data-src="images/cuda-thread-hieachy.png"/>
</div>
<aside class="notes">
<ul>
<li>Blokker er gruppert i et grid (1D, 2D, eller 3D).</li>
<li>….så vi har altså ett grid at blokker, hvor hver blokk inneholder mange tråder!</li>
<li>Hver blokk har en unik ID i gridet...</li>
<li>Alle blokker i et grid har samme dimensjon...</li>
</ul>
</aside>
</section>
<!--<section>
<h2>Thread hierarchy - Unique Ids</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Using thread ID and block ID we can make unique IDs for each thread. So that each thread can operate
on
different data items
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="500" data-src="images/cuda-thread-hieachy.png"/>
</div>
<aside class="notes">
<ul>
<li>Ved hjelp av tråd IDer og blokk IDer kan vi lage unike tråd IDer, slik at hver tråd kan
opererer
på forskjellige data elementer...
</li>
<li>Vi skal komme tilbake til dette straks...</li>
</ul>
</aside>
</section>-->
<section>
<h2>Thread hierarchy - Scalability</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Threads in the same block can
<ul>
<li>
Be synchronized
</li>
<li>
Share data (shared memory)
</li>
</ul>
</li>
<li>Threads in different blocks can not cooperate
<ul>
<li>Thread blocks are independent</li>
<li>Can be executed in any order -> scalability</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="500" data-src="images/cuda-thread-hieachy.png"/>
</div>
<aside class="notes">
<ul>
<li>Det at trådblokker er uavhengig, gjør at koden man skriver, kan skaleres med antall GPU
kjerner…
</li>
<li>slik at en GPU med flere kjerner vil kjøre programmet på raskere enn en GPU med færre
kjerner...
</li>
</ul>
<!--Backup:
Enabling programmers to write code that scales with the number of cores.
All threads in a grid execute the same kernel - Alle tråder i ett grid utfører den samme kernelen.-->
</aside>
</section>
<section>
<h2>How to launch a kernel?</h2>
<p>Execution configuration example</p>
<pre><code class="hljs" data-trim contenteditable>
dim3 dimGrid(2, 2); // 4 blocks (2D)
dim3 dimBlock(2, 4); // 8 threads per block (2D)
myKernel<<<dimGrid, dimBlock>>>();
</code></pre>
<p>dim3 is a structure with three properties, x, y and z.</p>
<img data-src="images/variables.png"/>
<!--<p>E.g. dimGrid.x, dimGrid.y, dimGrid.z</p>
<img height="200" data-src="images/launch-kernel2.png"/>-->
<aside class="notes">
Hvordan starter man en “kernel”? For å starte en kernel må man angi:
<ul>
<li>SEE SLIDE</li>
<li>...mellom trippel vinkelbraketter</li>
<li>Number of threads = 4 * 8 = 32 threads</li>
<li>Det er bare ett eksempel på hvordan man kan starte en kernel, vanligvis har man mange flere
tråder….
</li>
</ul>
<!--Backup:
The programming guide for NVIDIA CUDA states that the minimum number of blocks should be at
least twice the number of SMs in the device, preferably larger than 100, and that 64 is the
minimum of threads within a block.-->
</aside>
</section>
<!--<section>
<h2>How to determine unique thread IDs?</h2>
<p>Four useful built-in variables</p>
<table>
<thead>
<tr>
<th>Variable
</th>
<th>Type</th>
<th>Description</th>
</tr>
</thead>
<tbody>
<tr>
<td>gridDim</td>
<td>dim3</td>
<td>Dimensions of a grid</td>
</tr>
<tr>
<td>blockIdx</td>
<td>uint3</td>
<td>Block index within a grid</td>
</tr>
<tr>
<td>blockDim</td>
<td>dim3</td>
<td>Dimensions of a block</td>
</tr>
<tr>
<td>threadIdx</td>
<td>uint3</td>
<td>Thread index within a block</td>
</tr>
</tbody>
</table>
<img data-src="images/variables.png"/>
<aside class="notes">
De fire innebygde variablene er tilgjengelig inni en kernel....
</aside>
</section>-->
<section>
<h2>How to determine unique thread IDs?</h2>
<p>With this execution configuration</p>
<pre><code class="hljs" data-trim contenteditable>
dim3 dimGrid(3); // 3 blocks in 1D
dim3 dimBlock (5); // 5 threads per block in 1D
</code></pre>
<img height="300" data-src="images/cuda-grid-calculation.png"/>
<aside class="notes">
</aside>
</section>
<section>
<h2>How are kernels executed?</h2>
<ul>
<li>The GPU core is the stream processor (SP)
<ul>
<li>Able to run a single sequential thread</li>
</ul>
</li>
<li>SPs are grouped into streaming multiprocessors (SMs)
<ul>
<li>Can execute hundreds of threads concurrently</li>
<li>SMs is basically a SIMD processor</li>
</ul>
</li>
<li>There are multiple SMs per GPU</li>
</ul>
<img height="200" data-src="images/cuda-kernels-executed.png"/>
<aside class="notes">
<ul>
<li>
<li>For å forklarer hvordan GPUer utfører “kernels”, må vi først se på hvordan GPUer fra NVIDIA
er
designet ….
</li>
<li>Hver stream prosessor er i stand til å kjøre en sekvensiell tråd...
</li>
<li>Stream processorer er gruppert til streaming multiprocessorer...
</li>
<li>En streaming multiprosessor kan kjøre et stort antall tråder samtidig, hvor hver tråd kjører
det
samme program.
</li>
<li>En stream multiprosessor kan ses på som en SIMD-prosessor.</li>
</ul>
<!--Backup:
The multiprocessor creates, manages, schedules, and executes threads in groups of 32 parallel threads
called warps
An SM can execute large numbers of threads simultaneously, with each thread running the same program.
This SIMT (single instruction multiple thread) architecture is especially suitable for applications with
a high degree of data parallelism, where the same operations are applied to large amounts of data.
http://on-demand.gputechconf.com/gtc-express/2011/presentations/cuda_webinars_WarpsAndOccupancy.pdf-->
</aside>
</section>
<section>
<h2>NVIDIA's consumer graphics cards</h2>
<img height="500" data-src="images/cuda-overview.png"/>
<aside class="notes">
Consumer graphics cards: Forbruker grafikkort
</aside>
</section>
<section>
<h2>How are kernels executed - Grid</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Grid → GPU
<ul>
<li>An entire grid is handled by a single GPU chip</li>
<li>The GPU is responsible for allocating blocks to SMs that has available capacity</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="400" data-src="images/cuda-execution.png"/>
</div>
<aside class="notes">
<ul>
<li>Ett grid er håndert av en enkel GPU.
<li>Det er GPUen som er ansvarlig for allokering av blokker til streaming multiprocessors med
tilgjengelig kapasitet… Kan også nevne at dette blir gjort automatisk av GPUen...
</li>
</ul>
<!--URLs:
http://www.cc.gatech.edu/~vetter/keeneland/tutorial-2011-04-14/02-cuda-overview.pdf-->
</aside>
</section>
<section>
<h2>How are kernels executed - Block</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Block → Streaming multiprocessor
<ul>
<li>A block is never divided across multiple streaming multiprocessors</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="400" data-src="images/cuda-execution.png"/>
</div>
<aside class="notes">
<ul>
<li>En blokk blir håndert av bare en enkel streaming multiprocessor.</li>
</ul>
</aside>
</section>
<section>
<h2>How are kernels executed - Thread</h2>
<div class='left' style='float:left;width:48%'>
<ul>
<li>Thread → Stream processor
<ul>
<li>Each stream processor handles one or more threads in a block</li>
</ul>
</li>
</ul>
</div>
<div class='right' style='float:right;width:48%'>
<img height="400" data-src="images/cuda-execution.png"/>
</div>
<aside class="notes">
<ul>
<li>En stream processor kan håndtere en eller flere tråder i en blokk.</li>
</ul>
</aside>
</section>
<section>
<h2>Synchronization</h2>
<ul>
<li>Point in the program where threads stop and wait</li>
<li>When all threads have reached the barrier, they can proceed</li>
</ul>
<img height="200" data-src="images/cuda-sync.png"/>