Doc: Use overlays for some item lists
[cuda-ada.git] / doc / presentation.lyx
1 #LyX 2.0 created this file. For more info see http://www.lyx.org/
2 \lyxformat 413
3 \begin_document
4 \begin_header
5 \textclass beamer
6 \begin_preamble
7 \usepackage{listings}
8 \usetheme{Frankfurt}
9 % or ...
10 %\usetheme{Antibes}     % tree outline, neat
11 %\usetheme{JuanLesPins} % like Antibes, with shading
12 %\usetheme{Bergen}      % outline on side
13 %\usetheme{Luebeck}     % like Warsaw, square sides
14 %\usetheme{Berkeley}    % interesting left bar outline
15 %\usetheme{Madrid}      % clean, nice.  7/12 page numbers
16 %\usetheme{Berlin}      % dots show slide number
17 %\usetheme{Malmoe}      % OK, plain, unshaded
18 %\usetheme{Boadilla}    % nice, white bg, no top bar
19 %\usetheme{Marburg}     % nice, outline on right
20 %\usetheme{boxes}       % ???
21 %\usetheme{Montpellier} % tree outline on top, plainish white
22 %\usetheme{Copenhagen}  % like Warsaw
23 %\usetheme{PaloAlto}    % looks good
24 %\usetheme{Darmstadt}   % like Warsaw with circle outline
25 %\usetheme{Pittsburgh}
26 %\usetheme{default}
27 %\usetheme{Rochester}   % like boxy, unshaded warsaw
28 %\usetheme{Dresden}     % circle outline on top
29 %\usetheme{Singapore}   % purple gradient top
30 %\usetheme{Frankfurt}   % like Warsaw with circle outline on top
31 %\usetheme{Szeged}
32 %\usetheme{Goettingen}  % light purple right bar outline
33 %\usetheme{Warsaw}
34 %\usetheme{Hannover}    % like Goett with bar on left
35 %\usetheme{compatibility}
36 %\usetheme{Ilmenau}
37
38 \setbeamercovered{transparent}
39 % or whatever (possibly just delete it)
40
41 \useinnertheme{rectangles}
42
43 %\usecolortheme{seahorse}
44 %\usecolortheme{rose}
45
46 % seems to fix typewriter font in outline header:
47 \usepackage{ae,aecompl}
48
49 \definecolor{newyellow}{rgb}{1,1,0.8}
50 \definecolor{colKeys}{rgb}{0,0,1}
51 \definecolor{colIdentifier}{rgb}{0,0,0}
52 \definecolor{colComments}{rgb}{1,0,0}
53 \definecolor{colString}{rgb}{0,0.5,0}
54 \end_preamble
55 \use_default_options false
56 \maintain_unincluded_children false
57 \language english
58 \language_package default
59 \inputencoding auto
60 \fontencoding global
61 \font_roman default
62 \font_sans default
63 \font_typewriter default
64 \font_default_family default
65 \use_non_tex_fonts false
66 \font_sc false
67 \font_osf false
68 \font_sf_scale 100
69 \font_tt_scale 100
70
71 \graphics default
72 \default_output_format default
73 \output_sync 0
74 \bibtex_command default
75 \index_command default
76 \paperfontsize default
77 \spacing single
78 \use_hyperref false
79 \papersize default
80 \use_geometry true
81 \use_amsmath 2
82 \use_esint 0
83 \use_mhchem 1
84 \use_mathdots 1
85 \cite_engine basic
86 \use_bibtopic false
87 \use_indices false
88 \paperorientation portrait
89 \suppress_date false
90 \use_refstyle 0
91 \index Index
92 \shortcut idx
93 \color #008000
94 \end_index
95 \secnumdepth 2
96 \tocdepth 2
97 \paragraph_separation indent
98 \paragraph_indentation default
99 \quotes_language english
100 \papercolumns 1
101 \papersides 1
102 \paperpagestyle default
103 \listings_params "backgroundcolor={\color{newyellow}},basicstyle={\ttfamily\small},breakautoindent=true,breaklines=true,captionpos=b,commentstyle={\color{colComments}},extendedchars=true,frame=single,identifierstyle={\color{colIdentifier}},keywordstyle={\color{colKeys}},language=Ada,numbers=left,numberstyle={\tiny},showspaces=false,showstringspaces=false,stringstyle={\color{colString}},tabsize=4"
104 \tracking_changes false
105 \output_changes false
106 \html_math_output 0
107 \html_css_as_file 0
108 \html_be_strict false
109 \end_header
110
111 \begin_body
112
113 \begin_layout Title
114 CUDA/Ada
115 \begin_inset Argument
116 status open
117
118 \begin_layout Plain Layout
119 CUDA/Ada
120 \end_layout
121
122 \end_inset
123
124
125 \end_layout
126
127 \begin_layout Subtitle
128 An Ada binding to CUDA
129 \end_layout
130
131 \begin_layout Author
132 Reto Bürki, Adrian-Ken Rüegsegger
133 \begin_inset ERT
134 status collapsed
135
136 \begin_layout Plain Layout
137
138
139 \backslash
140
141 \backslash
142
143 \end_layout
144
145 \end_inset
146
147 University of Applied Sciences Rapperswil (HSR), Switzerland
148 \begin_inset Argument
149 status open
150
151 \begin_layout Plain Layout
152 Reto Bürki, Adrian-Ken Rüegsegger
153 \end_layout
154
155 \end_inset
156
157
158 \end_layout
159
160 \begin_layout Date
161 1/16/2012
162 \begin_inset ERT
163 status collapsed
164
165 \begin_layout Plain Layout
166
167
168 \backslash
169
170 \backslash
171
172 \end_layout
173
174 \end_inset
175
176 Master seminar: Progam Analysis and Transformation
177 \end_layout
178
179 \begin_layout BeginFrame
180 Outline
181 \end_layout
182
183 \begin_layout Standard
184 \begin_inset CommandInset toc
185 LatexCommand tableofcontents
186
187 \end_inset
188
189
190 \end_layout
191
192 \begin_layout EndFrame
193
194 \end_layout
195
196 \begin_layout Section
197 Introduction
198 \end_layout
199
200 \begin_layout Subsection
201 CUDA
202 \end_layout
203
204 \begin_layout BeginFrame
205 CUDA
206 \end_layout
207
208 \begin_layout Block
209 \begin_inset ERT
210 status collapsed
211
212 \begin_layout Plain Layout
213
214 {
215 \end_layout
216
217 \end_inset
218
219 What is CUDA?
220 \begin_inset ERT
221 status collapsed
222
223 \begin_layout Plain Layout
224
225 }
226 \end_layout
227
228 \end_inset
229
230
231 \end_layout
232
233 \begin_deeper
234 \begin_layout Itemize
235 Parallel computing architecture developed by NVIDIA
236 \end_layout
237
238 \begin_layout Itemize
239 \begin_inset Quotes eld
240 \end_inset
241
242 Compute Unified Device Architecture
243 \begin_inset Quotes erd
244 \end_inset
245
246
247 \end_layout
248
249 \begin_layout Itemize
250 General purpose computation using GPU (GPGPU)
251 \end_layout
252
253 \begin_layout Itemize
254 Use GPU as dedicated massively parallel co-processor
255 \end_layout
256
257 \begin_layout Itemize
258 Tremendous performance improvements possible
259 \end_layout
260
261 \end_deeper
262 \begin_layout EndFrame
263
264 \end_layout
265
266 \begin_layout BeginFrame
267 CUDA Processing
268 \end_layout
269
270 \begin_layout Standard
271 \align center
272 \begin_inset Graphics
273         filename cuda-processing.png
274         scale 33
275
276 \end_inset
277
278
279 \end_layout
280
281 \begin_layout EndFrame
282
283 \end_layout
284
285 \begin_layout Subsection
286 Ada
287 \end_layout
288
289 \begin_layout BeginFrame
290 Ada
291 \end_layout
292
293 \begin_layout Block
294 \begin_inset ERT
295 status collapsed
296
297 \begin_layout Plain Layout
298
299 {
300 \end_layout
301
302 \end_inset
303
304 The Ada programming language
305 \begin_inset ERT
306 status collapsed
307
308 \begin_layout Plain Layout
309
310 }
311 \end_layout
312
313 \end_inset
314
315
316 \end_layout
317
318 \begin_deeper
319 \begin_layout Itemize
320 \begin_inset ERT
321 status collapsed
322
323 \begin_layout Plain Layout
324
325 <1->
326 \end_layout
327
328 \end_inset
329
330 Structured, strongly typed programming language
331 \end_layout
332
333 \begin_layout Itemize
334 \begin_inset ERT
335 status collapsed
336
337 \begin_layout Plain Layout
338
339 <2->
340 \end_layout
341
342 \end_inset
343
344 Initiated by the US Department of Defense (DoD)
345 \end_layout
346
347 \begin_layout Itemize
348 \begin_inset ERT
349 status collapsed
350
351 \begin_layout Plain Layout
352
353 <3->
354 \end_layout
355
356 \end_inset
357
358 First standardized high-level programming language
359 \end_layout
360
361 \begin_layout Itemize
362 \begin_inset ERT
363 status collapsed
364
365 \begin_layout Plain Layout
366
367 <4->
368 \end_layout
369
370 \end_inset
371
372 Emphasizes safety and security
373 \end_layout
374
375 \begin_layout Itemize
376 \begin_inset ERT
377 status collapsed
378
379 \begin_layout Plain Layout
380
381 <5->
382 \end_layout
383
384 \end_inset
385
386 Supports all modern programming paradigms
387 \end_layout
388
389 \begin_layout Itemize
390 \begin_inset ERT
391 status collapsed
392
393 \begin_layout Plain Layout
394
395 <6->
396 \end_layout
397
398 \end_inset
399
400 Current language standard is Ada 2005, next release 2012
401 \end_layout
402
403 \begin_layout Itemize
404 \begin_inset ERT
405 status collapsed
406
407 \begin_layout Plain Layout
408
409 <7->
410 \end_layout
411
412 \end_inset
413
414 Mostly used in aviation, railway systems, banking, military and space technology
415 \end_layout
416
417 \end_deeper
418 \begin_layout EndFrame
419
420 \end_layout
421
422 \begin_layout BeginFrame
423 Ada Compiler
424 \end_layout
425
426 \begin_layout Block
427 \begin_inset ERT
428 status collapsed
429
430 \begin_layout Plain Layout
431
432 {
433 \end_layout
434
435 \end_inset
436
437 GNAT
438 \begin_inset ERT
439 status collapsed
440
441 \begin_layout Plain Layout
442
443 }
444 \end_layout
445
446 \end_inset
447
448
449 \end_layout
450
451 \begin_deeper
452 \begin_layout Itemize
453 \begin_inset ERT
454 status collapsed
455
456 \begin_layout Plain Layout
457
458 <1->
459 \end_layout
460
461 \end_inset
462
463 Free-software compiler for Ada
464 \end_layout
465
466 \begin_layout Itemize
467 \begin_inset ERT
468 status collapsed
469
470 \begin_layout Plain Layout
471
472 <2->
473 \end_layout
474
475 \end_inset
476
477 Part of the GNU Compiler Collection (GCC)
478 \end_layout
479
480 \begin_layout Itemize
481 \begin_inset ERT
482 status collapsed
483
484 \begin_layout Plain Layout
485
486 <3->
487 \end_layout
488
489 \end_inset
490
491 Supports all versions of Ada (83, 95, 2005)
492 \end_layout
493
494 \begin_layout Itemize
495 \begin_inset ERT
496 status collapsed
497
498 \begin_layout Plain Layout
499
500 <4->
501 \end_layout
502
503 \end_inset
504
505 Available on most operating systems
506 \end_layout
507
508 \begin_layout Itemize
509 \begin_inset ERT
510 status collapsed
511
512 \begin_layout Plain Layout
513
514 <5->
515 \end_layout
516
517 \end_inset
518
519 100% compliant with Ada Conformity Assessment Test Suite (ACATS)
520 \end_layout
521
522 \end_deeper
523 \begin_layout EndFrame
524
525 \end_layout
526
527 \begin_layout Subsection
528 Motivation
529 \end_layout
530
531 \begin_layout BeginFrame
532 Motivation
533 \end_layout
534
535 \begin_layout Block
536 \begin_inset ERT
537 status collapsed
538
539 \begin_layout Plain Layout
540
541 {
542 \end_layout
543
544 \end_inset
545
546 Why CUDA/Ada?
547 \begin_inset ERT
548 status collapsed
549
550 \begin_layout Plain Layout
551
552 }
553 \end_layout
554
555 \end_inset
556
557
558 \end_layout
559
560 \begin_deeper
561 \begin_layout Itemize
562 \begin_inset ERT
563 status collapsed
564
565 \begin_layout Plain Layout
566
567 <1->
568 \end_layout
569
570 \end_inset
571
572 CUDA wrappers exist for many languages but not for Ada
573 \end_layout
574
575 \begin_layout Itemize
576 \begin_inset ERT
577 status collapsed
578
579 \begin_layout Plain Layout
580
581 <2->
582 \end_layout
583
584 \end_inset
585
586 Make CUDA accessible for Ada developers
587 \end_layout
588
589 \begin_layout Itemize
590 \begin_inset ERT
591 status collapsed
592
593 \begin_layout Plain Layout
594
595 <3->
596 \end_layout
597
598 \end_inset
599
600 Benefit from the advantages and processing power of a GPU in Ada applications
601 \end_layout
602
603 \begin_layout Itemize
604 \begin_inset ERT
605 status collapsed
606
607 \begin_layout Plain Layout
608
609 <4->
610 \end_layout
611
612 \end_inset
613
614 Curiosity
615 \end_layout
616
617 \begin_deeper
618 \begin_layout Itemize
619 How well do CUDA and Ada match up?
620 \end_layout
621
622 \begin_layout Itemize
623 Can it be done in a nice way?
624 \end_layout
625
626 \begin_layout Itemize
627 Possible perfomance penalties from Ada runtime?
628 \end_layout
629
630 \end_deeper
631 \end_deeper
632 \begin_layout EndFrame
633
634 \end_layout
635
636 \begin_layout Section
637 Bindings in Ada
638 \end_layout
639
640 \begin_layout Subsection
641 pragma Import
642 \end_layout
643
644 \begin_layout BeginFrame
645 Bindings in Ada
646 \end_layout
647
648 \begin_layout Block
649 \begin_inset ERT
650 status collapsed
651
652 \begin_layout Plain Layout
653
654 {
655 \end_layout
656
657 \end_inset
658
659
660 \family typewriter
661 Import
662 \family default
663  pragma
664 \begin_inset ERT
665 status collapsed
666
667 \begin_layout Plain Layout
668
669 }
670 \end_layout
671
672 \end_inset
673
674
675 \end_layout
676
677 \begin_deeper
678 \begin_layout Itemize
679 \begin_inset ERT
680 status collapsed
681
682 \begin_layout Plain Layout
683
684 <1->
685 \end_layout
686
687 \end_inset
688
689 Used to access functionality which is written in another programming language
690 \end_layout
691
692 \begin_layout Itemize
693 \begin_inset ERT
694 status collapsed
695
696 \begin_layout Plain Layout
697
698 <2->
699 \end_layout
700
701 \end_inset
702
703 Pragmas are directives which control the compiler
704 \end_layout
705
706 \begin_layout Itemize
707 \begin_inset ERT
708 status collapsed
709
710 \begin_layout Plain Layout
711
712 <3->
713 \end_layout
714
715 \end_inset
716
717 General form: 
718 \family typewriter
719 pragma Name (Parameter_List);
720 \end_layout
721
722 \begin_layout Itemize
723 \begin_inset ERT
724 status collapsed
725
726 \begin_layout Plain Layout
727
728 <4->
729 \end_layout
730
731 \end_inset
732
733 Predefined packages that make it easier to interface with C
734 \end_layout
735
736 \begin_deeper
737 \begin_layout Itemize
738
739 \family typewriter
740 Interfaces.C
741 \end_layout
742
743 \begin_layout Itemize
744
745 \family typewriter
746 System
747 \end_layout
748
749 \end_deeper
750 \end_deeper
751 \begin_layout EndFrame
752
753 \end_layout
754
755 \begin_layout Standard
756 \begin_inset ERT
757 status open
758
759 \begin_layout Plain Layout
760
761
762 \backslash
763 begin{frame}[fragile]
764 \backslash
765 frametitle{Importing a C function}
766 \end_layout
767
768 \end_inset
769
770
771 \end_layout
772
773 \begin_layout Standard
774
775 \family typewriter
776 \series bold
777 C function
778 \end_layout
779
780 \begin_layout Standard
781 \begin_inset listings
782 lstparams "language=C"
783 inline false
784 status open
785
786 \begin_layout Plain Layout
787
788 int bind(int sockfd, const struct sockaddr *addr, socklen_t addrlen);
789 \end_layout
790
791 \end_inset
792
793
794 \end_layout
795
796 \begin_layout Standard
797
798 \family typewriter
799 \series bold
800 Ada import
801 \end_layout
802
803 \begin_layout Standard
804 \begin_inset listings
805 inline false
806 status open
807
808 \begin_layout Plain Layout
809
810 function C_Bind
811 \end_layout
812
813 \begin_layout Plain Layout
814
815  (S       : Interfaces.C.int;
816 \end_layout
817
818 \begin_layout Plain Layout
819
820   Name    : System.Address;
821 \end_layout
822
823 \begin_layout Plain Layout
824
825   Namelen : Interfaces.C.int)
826 \end_layout
827
828 \begin_layout Plain Layout
829
830   return Interfaces.C.int; 
831 \end_layout
832
833 \begin_layout Plain Layout
834
835 pragma Import (C, C_Bind, "bind"); 
836 \end_layout
837
838 \end_inset
839
840
841 \end_layout
842
843 \begin_layout Standard
844 \begin_inset ERT
845 status open
846
847 \begin_layout Plain Layout
848
849
850 \backslash
851 end{frame}
852 \end_layout
853
854 \end_inset
855
856
857 \end_layout
858
859 \begin_layout Subsection
860 Definition
861 \end_layout
862
863 \begin_layout BeginFrame
864
865 \family typewriter
866 Definition
867 \end_layout
868
869 \begin_layout AlertBlock
870 \begin_inset ERT
871 status collapsed
872
873 \begin_layout Plain Layout
874
875 {
876 \end_layout
877
878 \end_inset
879
880
881 \family typewriter
882 Binding
883 \family default
884
885 \begin_inset ERT
886 status collapsed
887
888 \begin_layout Plain Layout
889
890 }
891 \end_layout
892
893 \end_inset
894
895
896 \end_layout
897
898 \begin_deeper
899 \begin_layout Quote
900 A library which wraps another library not written in Ada is called a 
901 \begin_inset Quotes eld
902 \end_inset
903
904 binding
905 \begin_inset Quotes erd
906 \end_inset
907
908 .
909  Other programming languages also use the term 
910 \begin_inset Quotes eld
911 \end_inset
912
913 wrapper
914 \begin_inset Quotes erd
915 \end_inset
916
917  or 
918 \begin_inset Quotes eld
919 \end_inset
920
921 library wrapper
922 \begin_inset Quotes erd
923 \end_inset
924
925 .
926  
927 \end_layout
928
929 \end_deeper
930 \begin_layout EndFrame
931
932 \end_layout
933
934 \begin_layout Subsection
935 Thin- and Thick-Binding
936 \end_layout
937
938 \begin_layout BeginFrame
939 Thin-Binding
940 \end_layout
941
942 \begin_layout Block
943 \begin_inset ERT
944 status collapsed
945
946 \begin_layout Plain Layout
947
948 {
949 \end_layout
950
951 \end_inset
952
953
954 \family typewriter
955 What is a Thin-Binding
956 \family default
957 ?
958 \begin_inset ERT
959 status collapsed
960
961 \begin_layout Plain Layout
962
963 }
964 \end_layout
965
966 \end_inset
967
968
969 \end_layout
970
971 \begin_deeper
972 \begin_layout Itemize
973 One-to-one mapping of the foreign library interface to Ada
974 \end_layout
975
976 \begin_layout Itemize
977 Straight forward to create but time consuming and error prone
978 \end_layout
979
980 \begin_layout Itemize
981 Cumbersome to work with (because of direct mapping)
982 \end_layout
983
984 \begin_layout Itemize
985 No protection normally guaranteed by Ada
986 \end_layout
987
988 \end_deeper
989 \begin_layout EndFrame
990
991 \end_layout
992
993 \begin_layout BeginFrame
994 Thick-Binding
995 \end_layout
996
997 \begin_layout Block
998 \begin_inset ERT
999 status collapsed
1000
1001 \begin_layout Plain Layout
1002
1003 {
1004 \end_layout
1005
1006 \end_inset
1007
1008
1009 \family typewriter
1010 What is a Thick-Binding
1011 \family default
1012 ?
1013 \begin_inset ERT
1014 status collapsed
1015
1016 \begin_layout Plain Layout
1017
1018 }
1019 \end_layout
1020
1021 \end_inset
1022
1023
1024 \end_layout
1025
1026 \begin_deeper
1027 \begin_layout Itemize
1028 Provides a more abstract, Ada-like view of foreign library
1029 \end_layout
1030
1031 \begin_layout Itemize
1032 Provides proper Ada types and operations
1033 \end_layout
1034
1035 \begin_layout Itemize
1036 Ensure safe usage of wrapper library
1037 \end_layout
1038
1039 \begin_layout Itemize
1040 Easier to work with but takes more work and time to create
1041 \end_layout
1042
1043 \end_deeper
1044 \begin_layout EndFrame
1045
1046 \end_layout
1047
1048 \begin_layout BeginFrame
1049 Thin- and Thick-Bindings
1050 \end_layout
1051
1052 \begin_layout AlertBlock
1053 \begin_inset ERT
1054 status collapsed
1055
1056 \begin_layout Plain Layout
1057
1058 {
1059 \end_layout
1060
1061 \end_inset
1062
1063 Using both
1064 \begin_inset ERT
1065 status collapsed
1066
1067 \begin_layout Plain Layout
1068
1069 }
1070 \end_layout
1071
1072 \end_inset
1073
1074
1075 \end_layout
1076
1077 \begin_deeper
1078 \begin_layout Itemize
1079 Thin- and Thick-Binding layers are often used in conjunction
1080 \end_layout
1081
1082 \begin_layout Itemize
1083 Thin-Binding wraps foreign language library (low-level)
1084 \end_layout
1085
1086 \begin_layout Itemize
1087 Thick-Binding abstracts Thin-Binding to provide an Ada-like 
1088 \begin_inset Quotes eld
1089 \end_inset
1090
1091 look and feel
1092 \begin_inset Quotes erd
1093 \end_inset
1094
1095
1096 \end_layout
1097
1098 \begin_layout Itemize
1099 Separation improves maintainability because both layers can be adapted when
1100  needed
1101 \end_layout
1102
1103 \end_deeper
1104 \begin_layout EndFrame
1105
1106 \end_layout
1107
1108 \begin_layout Section
1109 CUDA/Ada Design
1110 \end_layout
1111
1112 \begin_layout Subsection
1113 Design Goals
1114 \end_layout
1115
1116 \begin_layout BeginFrame
1117 Design Goals
1118 \end_layout
1119
1120 \begin_layout Block
1121 \begin_inset ERT
1122 status collapsed
1123
1124 \begin_layout Plain Layout
1125
1126 {
1127 \end_layout
1128
1129 \end_inset
1130
1131
1132 \family typewriter
1133 Inspiration
1134 \family default
1135
1136 \begin_inset ERT
1137 status collapsed
1138
1139 \begin_layout Plain Layout
1140
1141 }
1142 \end_layout
1143
1144 \end_inset
1145
1146
1147 \end_layout
1148
1149 \begin_deeper
1150 \begin_layout Itemize
1151 CUDA/Ada heavily inspired by PyCUDA
1152 \end_layout
1153
1154 \begin_layout Itemize
1155 Great binding for Python from Andreas Klöckner
1156 \end_layout
1157
1158 \end_deeper
1159 \begin_layout EndFrame
1160
1161 \end_layout
1162
1163 \begin_layout BeginFrame
1164 Design Goals II
1165 \end_layout
1166
1167 \begin_layout Block
1168 \begin_inset ERT
1169 status collapsed
1170
1171 \begin_layout Plain Layout
1172
1173 {
1174 \end_layout
1175
1176 \end_inset
1177
1178
1179 \family typewriter
1180 Goals
1181 \family default
1182
1183 \begin_inset ERT
1184 status collapsed
1185
1186 \begin_layout Plain Layout
1187
1188 }
1189 \end_layout
1190
1191 \end_inset
1192
1193
1194 \end_layout
1195
1196 \begin_deeper
1197 \begin_layout Itemize
1198 \begin_inset ERT
1199 status collapsed
1200
1201 \begin_layout Plain Layout
1202
1203 <1->
1204 \end_layout
1205
1206 \end_inset
1207
1208 Seamless access to CUDA from Ada
1209 \end_layout
1210
1211 \begin_layout Itemize
1212 \begin_inset ERT
1213 status collapsed
1214
1215 \begin_layout Plain Layout
1216
1217 <2->
1218 \end_layout
1219
1220 \end_inset
1221
1222 High abstraction
1223 \end_layout
1224
1225 \begin_layout Itemize
1226 \begin_inset ERT
1227 status collapsed
1228
1229 \begin_layout Plain Layout
1230
1231 <3->
1232 \end_layout
1233
1234 \end_inset
1235
1236 Auto-initialization
1237 \end_layout
1238
1239 \begin_layout Itemize
1240 \begin_inset ERT
1241 status collapsed
1242
1243 \begin_layout Plain Layout
1244
1245 <4->
1246 \end_layout
1247
1248 \end_inset
1249
1250 JIT-Compilation of CUDA kernels
1251 \end_layout
1252
1253 \begin_layout Itemize
1254 \begin_inset ERT
1255 status collapsed
1256
1257 \begin_layout Plain Layout
1258
1259 <5->
1260 \end_layout
1261
1262 \end_inset
1263
1264 Convenient argument handling
1265 \end_layout
1266
1267 \begin_layout Itemize
1268 \begin_inset ERT
1269 status collapsed
1270
1271 \begin_layout Plain Layout
1272
1273 <6->
1274 \end_layout
1275
1276 \end_inset
1277
1278 Error-handling using Ada Exceptions
1279 \end_layout
1280
1281 \begin_layout Itemize
1282 \begin_inset ERT
1283 status collapsed
1284
1285 \begin_layout Plain Layout
1286
1287 <7->
1288 \end_layout
1289
1290 \end_inset
1291
1292 Speed
1293 \end_layout
1294
1295 \begin_layout Itemize
1296 \begin_inset ERT
1297 status collapsed
1298
1299 \begin_layout Plain Layout
1300
1301 <8->
1302 \end_layout
1303
1304 \end_inset
1305
1306 Automatically generated Thin-Binding
1307 \end_layout
1308
1309 \end_deeper
1310 \begin_layout EndFrame
1311
1312 \end_layout
1313
1314 \begin_layout Section
1315 CUDA Binding
1316 \end_layout
1317
1318 \begin_layout Subsection
1319 Thin-Binding
1320 \end_layout
1321
1322 \begin_layout Standard
1323 \begin_inset ERT
1324 status open
1325
1326 \begin_layout Plain Layout
1327
1328
1329 \backslash
1330 begin{frame}[fragile]
1331 \backslash
1332 frametitle{Thin-Binding to CUDA}
1333 \end_layout
1334
1335 \end_inset
1336
1337
1338 \end_layout
1339
1340 \begin_layout Block
1341 \begin_inset ERT
1342 status collapsed
1343
1344 \begin_layout Plain Layout
1345
1346 {
1347 \end_layout
1348
1349 \end_inset
1350
1351
1352 \family typewriter
1353 Creation
1354 \family default
1355
1356 \begin_inset ERT
1357 status collapsed
1358
1359 \begin_layout Plain Layout
1360
1361 }
1362 \end_layout
1363
1364 \end_inset
1365
1366
1367 \end_layout
1368
1369 \begin_deeper
1370 \begin_layout Itemize
1371 Auto-generated using 
1372 \family typewriter
1373 -fdump-ada-spec
1374 \family default
1375  of GNAT
1376 \end_layout
1377
1378 \begin_layout Itemize
1379 CUDA/Ada imports CUDA driver and runtime API 
1380 \end_layout
1381
1382 \begin_deeper
1383 \begin_layout Itemize
1384
1385 \family typewriter
1386 cuda.h
1387 \end_layout
1388
1389 \begin_layout Itemize
1390
1391 \family typewriter
1392 cuda_runtime.h
1393 \end_layout
1394
1395 \end_deeper
1396 \begin_layout Itemize
1397 Support for i686 and x86_64
1398 \end_layout
1399
1400 \begin_deeper
1401 \begin_layout Itemize
1402 e.g.
1403  CUdeviceptr:
1404 \family typewriter
1405 unsigned
1406 \family default
1407  vs.
1408
1409 \family typewriter
1410 unsigned_long_long
1411 \end_layout
1412
1413 \end_deeper
1414 \end_deeper
1415 \begin_layout Standard
1416 \begin_inset ERT
1417 status open
1418
1419 \begin_layout Plain Layout
1420
1421
1422 \backslash
1423 begin{verbatim}
1424 \end_layout
1425
1426 \begin_layout Plain Layout
1427
1428 gcc -c -fdump-ada-spec cuda.h
1429 \end_layout
1430
1431 \begin_layout Plain Layout
1432
1433 gcc -c -fdump-ada-spec cuda_runtime.h
1434 \end_layout
1435
1436 \begin_layout Plain Layout
1437
1438
1439 \backslash
1440 end{verbatim}
1441 \end_layout
1442
1443 \end_inset
1444
1445
1446 \end_layout
1447
1448 \begin_layout Standard
1449 \begin_inset ERT
1450 status open
1451
1452 \begin_layout Plain Layout
1453
1454
1455 \backslash
1456 end{frame}
1457 \end_layout
1458
1459 \end_inset
1460
1461
1462 \end_layout
1463
1464 \begin_layout Standard
1465 \begin_inset ERT
1466 status open
1467
1468 \begin_layout Plain Layout
1469
1470
1471 \backslash
1472 begin{frame}[fragile]
1473 \backslash
1474 frametitle{Architecture (build logic)}
1475 \end_layout
1476
1477 \end_inset
1478
1479
1480 \end_layout
1481
1482 \begin_layout Standard
1483 \begin_inset ERT
1484 status open
1485
1486 \begin_layout Plain Layout
1487
1488
1489 \backslash
1490 begin{verbatim}
1491 \end_layout
1492
1493 \begin_layout Plain Layout
1494
1495 ARCH ?= $(shell uname -m)
1496 \end_layout
1497
1498 \begin_layout Plain Layout
1499
1500 gnatmake -Pcuda -XARCH=$(ARCH)
1501 \end_layout
1502
1503 \begin_layout Plain Layout
1504
1505
1506 \backslash
1507 end{verbatim}
1508 \end_layout
1509
1510 \end_inset
1511
1512
1513 \end_layout
1514
1515 \begin_layout Standard
1516 \begin_inset listings
1517 inline false
1518 status open
1519
1520 \begin_layout Plain Layout
1521
1522 with "thin/binding";
1523 \end_layout
1524
1525 \end_inset
1526
1527
1528 \end_layout
1529
1530 \begin_layout Standard
1531 \begin_inset listings
1532 inline false
1533 status open
1534
1535 \begin_layout Plain Layout
1536
1537 type Arch_Type is ("x86_64", "i686"); 
1538 \end_layout
1539
1540 \begin_layout Plain Layout
1541
1542 Arch : Arch_Type := external ("ARCH", "x86_64");
1543 \end_layout
1544
1545 \begin_layout Plain Layout
1546
1547 \end_layout
1548
1549 \begin_layout Plain Layout
1550
1551 for Source_Dirs use (".", ARCH);
1552 \end_layout
1553
1554 \end_inset
1555
1556
1557 \end_layout
1558
1559 \begin_layout Standard
1560 \begin_inset ERT
1561 status open
1562
1563 \begin_layout Plain Layout
1564
1565
1566 \backslash
1567 end{frame}
1568 \end_layout
1569
1570 \end_inset
1571
1572
1573 \end_layout
1574
1575 \begin_layout Subsection
1576 Thick-Binding
1577 \end_layout
1578
1579 \begin_layout Standard
1580 \begin_inset ERT
1581 status open
1582
1583 \begin_layout Plain Layout
1584
1585
1586 \backslash
1587 begin{frame}[fragile]
1588 \backslash
1589 frametitle{Autoinit}
1590 \end_layout
1591
1592 \end_inset
1593
1594
1595 \end_layout
1596
1597 \begin_layout Block
1598 \begin_inset ERT
1599 status collapsed
1600
1601 \begin_layout Plain Layout
1602
1603 {
1604 \end_layout
1605
1606 \end_inset
1607
1608
1609 \family typewriter
1610 Functionality
1611 \family default
1612
1613 \begin_inset ERT
1614 status collapsed
1615
1616 \begin_layout Plain Layout
1617
1618 }
1619 \end_layout
1620
1621 \end_inset
1622
1623
1624 \end_layout
1625
1626 \begin_deeper
1627 \begin_layout Itemize
1628 Takes care of CUDA initialisation task
1629 \end_layout
1630
1631 \begin_deeper
1632 \begin_layout Itemize
1633 cuInit
1634 \end_layout
1635
1636 \begin_layout Itemize
1637 cuDeviceGet
1638 \end_layout
1639
1640 \begin_layout Itemize
1641 cuCtxCreate
1642 \end_layout
1643
1644 \end_deeper
1645 \begin_layout Itemize
1646 Handles release of CUDA context
1647 \end_layout
1648
1649 \begin_deeper
1650 \begin_layout Itemize
1651 cuCtxDestroy
1652 \end_layout
1653
1654 \end_deeper
1655 \end_deeper
1656 \begin_layout Standard
1657 \begin_inset listings
1658 inline false
1659 status open
1660
1661 \begin_layout Plain Layout
1662
1663 with CUDA.Autoinit;
1664 \end_layout
1665
1666 \end_inset
1667
1668
1669 \end_layout
1670
1671 \begin_layout Standard
1672 \begin_inset ERT
1673 status open
1674
1675 \begin_layout Plain Layout
1676
1677
1678 \backslash
1679 end{frame}
1680 \end_layout
1681
1682 \end_inset
1683
1684
1685 \end_layout
1686
1687 \begin_layout Standard
1688 \begin_inset ERT
1689 status open
1690
1691 \begin_layout Plain Layout
1692
1693
1694 \backslash
1695 begin{frame}[fragile]
1696 \backslash
1697 frametitle{Source-Modules}
1698 \end_layout
1699
1700 \end_inset
1701
1702
1703 \end_layout
1704
1705 \begin_layout Block
1706 \begin_inset ERT
1707 status collapsed
1708
1709 \begin_layout Plain Layout
1710
1711 {
1712 \end_layout
1713
1714 \end_inset
1715
1716
1717 \family typewriter
1718 Functionality
1719 \family default
1720
1721 \begin_inset ERT
1722 status collapsed
1723
1724 \begin_layout Plain Layout
1725
1726 }
1727 \end_layout
1728
1729 \end_inset
1730
1731
1732 \end_layout
1733
1734 \begin_deeper
1735 \begin_layout Itemize
1736 Used to define CUDA kernels 
1737 \begin_inset Quotes eld
1738 \end_inset
1739
1740 inline
1741 \begin_inset Quotes erd
1742 \end_inset
1743
1744
1745 \end_layout
1746
1747 \end_deeper
1748 \begin_layout Standard
1749 \begin_inset listings
1750 inline false
1751 status open
1752
1753 \begin_layout Plain Layout
1754
1755 N   : constant                    := 32 * 1024;
1756 \end_layout
1757
1758 \begin_layout Plain Layout
1759
1760 Src : Compiler.Source_Module_Type :=
1761 \end_layout
1762
1763 \begin_layout Plain Layout
1764
1765  Compiler.Create
1766 \end_layout
1767
1768 \begin_layout Plain Layout
1769
1770   (Preamble  => "#define N" & N'Img,
1771 \end_layout
1772
1773 \begin_layout Plain Layout
1774
1775    Operation => "__global__ void add(float *a, float *b) {" 
1776 \end_layout
1777
1778 \begin_layout Plain Layout
1779
1780     & "  int tid = blockIdx.x;"                         
1781 \end_layout
1782
1783 \begin_layout Plain Layout
1784
1785     & "  while (tid < N) {"
1786 \end_layout
1787
1788 \begin_layout Plain Layout
1789
1790     & "    b[tid] = a[tid] + 10;"                       
1791 \end_layout
1792
1793 \begin_layout Plain Layout
1794
1795     & "    tid += gridDim.x;"
1796 \end_layout
1797
1798 \begin_layout Plain Layout
1799
1800     & "}}");
1801 \end_layout
1802
1803 \end_inset
1804
1805
1806 \end_layout
1807
1808 \begin_layout Standard
1809 \begin_inset ERT
1810 status open
1811
1812 \begin_layout Plain Layout
1813
1814
1815 \backslash
1816 end{frame}
1817 \end_layout
1818
1819 \end_inset
1820
1821
1822 \end_layout
1823
1824 \begin_layout Standard
1825 \begin_inset ERT
1826 status open
1827
1828 \begin_layout Plain Layout
1829
1830
1831 \backslash
1832 begin{frame}[fragile]
1833 \backslash
1834 frametitle{JIT-Compiler}
1835 \end_layout
1836
1837 \end_inset
1838
1839
1840 \end_layout
1841
1842 \begin_layout Block
1843 \begin_inset ERT
1844 status collapsed
1845
1846 \begin_layout Plain Layout
1847
1848 {
1849 \end_layout
1850
1851 \end_inset
1852
1853
1854 \family typewriter
1855 Functionality
1856 \family default
1857
1858 \begin_inset ERT
1859 status collapsed
1860
1861 \begin_layout Plain Layout
1862
1863 }
1864 \end_layout
1865
1866 \end_inset
1867
1868
1869 \end_layout
1870
1871 \begin_deeper
1872 \begin_layout Itemize
1873 Compile CUDA kernels at runtime
1874 \end_layout
1875
1876 \begin_layout Itemize
1877 Uses 
1878 \family typewriter
1879 nvcc
1880 \family default
1881  to generate CUBIN binary code
1882 \end_layout
1883
1884 \begin_layout Itemize
1885 Upload modules to GPU
1886 \end_layout
1887
1888 \begin_layout Itemize
1889 Caching of compiled modules
1890 \end_layout
1891
1892 \end_deeper
1893 \begin_layout Standard
1894 \begin_inset listings
1895 inline false
1896 status open
1897
1898 \begin_layout Plain Layout
1899
1900    ...
1901 \end_layout
1902
1903 \begin_layout Plain Layout
1904
1905 is
1906 \end_layout
1907
1908 \begin_layout Plain Layout
1909
1910    Module : Compiler.Module_Type; 
1911 \end_layout
1912
1913 \begin_layout Plain Layout
1914
1915 begin
1916 \end_layout
1917
1918 \begin_layout Plain Layout
1919
1920    Module := Compiler.Compile (Source => Src); 
1921 \end_layout
1922
1923 \begin_layout Plain Layout
1924
1925    ...
1926  
1927 \end_layout
1928
1929 \end_inset
1930
1931
1932 \end_layout
1933
1934 \begin_layout Standard
1935 \begin_inset ERT
1936 status open
1937
1938 \begin_layout Plain Layout
1939
1940
1941 \backslash
1942 end{frame}
1943 \end_layout
1944
1945 \end_inset
1946
1947
1948 \end_layout
1949
1950 \begin_layout BeginFrame
1951 JIT Workflow
1952 \end_layout
1953
1954 \begin_layout Standard
1955 \align center
1956 \begin_inset Graphics
1957         filename jit-compiler.eps
1958         scale 53
1959
1960 \end_inset
1961
1962
1963 \end_layout
1964
1965 \begin_layout EndFrame
1966
1967 \end_layout
1968
1969 \begin_layout Standard
1970 \begin_inset ERT
1971 status open
1972
1973 \begin_layout Plain Layout
1974
1975
1976 \backslash
1977 begin{frame}[fragile]
1978 \backslash
1979 frametitle{Calling a function}
1980 \end_layout
1981
1982 \end_inset
1983
1984
1985 \end_layout
1986
1987 \begin_layout Standard
1988 \begin_inset listings
1989 inline false
1990 status open
1991
1992 \begin_layout Plain Layout
1993
1994    ...
1995 \end_layout
1996
1997 \begin_layout Plain Layout
1998
1999    Func   : Compiler.Function_Type;
2000 \end_layout
2001
2002 \begin_layout Plain Layout
2003
2004    Module : Compiler.Module_Type; 
2005 \end_layout
2006
2007 \begin_layout Plain Layout
2008
2009 begin
2010 \end_layout
2011
2012 \begin_layout Plain Layout
2013
2014    Module := Compiler.Compile (Source => Src); 
2015 \end_layout
2016
2017 \begin_layout Plain Layout
2018
2019    Func   := Compiler.Get_Function
2020 \end_layout
2021
2022 \begin_layout Plain Layout
2023
2024      (Module => Module,
2025 \end_layout
2026
2027 \begin_layout Plain Layout
2028
2029       Name   => "add");
2030 \end_layout
2031
2032 \begin_layout Plain Layout
2033
2034 \end_layout
2035
2036 \begin_layout Plain Layout
2037
2038    Func.Call 
2039 \end_layout
2040
2041 \begin_layout Plain Layout
2042
2043      (Args =>
2044 \end_layout
2045
2046 \begin_layout Plain Layout
2047
2048         (1 => In_Arg (Data => A),
2049 \end_layout
2050
2051 \begin_layout Plain Layout
2052
2053          2 => In_Arg (Data => B),
2054 \end_layout
2055
2056 \begin_layout Plain Layout
2057
2058          3 => Out_Arg (Data => C'Access)));
2059 \end_layout
2060
2061 \end_inset
2062
2063
2064 \end_layout
2065
2066 \begin_layout Standard
2067 \begin_inset ERT
2068 status open
2069
2070 \begin_layout Plain Layout
2071
2072
2073 \backslash
2074 end{frame}
2075 \end_layout
2076
2077 \end_inset
2078
2079
2080 \end_layout
2081
2082 \begin_layout BeginFrame
2083 Kernel arguments
2084 \end_layout
2085
2086 \begin_layout Block
2087 \begin_inset ERT
2088 status collapsed
2089
2090 \begin_layout Plain Layout
2091
2092 {
2093 \end_layout
2094
2095 \end_inset
2096
2097
2098 \family typewriter
2099 Functionality
2100 \family default
2101
2102 \begin_inset ERT
2103 status collapsed
2104
2105 \begin_layout Plain Layout
2106
2107 }
2108 \end_layout
2109
2110 \end_inset
2111
2112
2113 \end_layout
2114
2115 \begin_deeper
2116 \begin_layout Itemize
2117 \begin_inset ERT
2118 status collapsed
2119
2120 \begin_layout Plain Layout
2121
2122 <1->
2123 \end_layout
2124
2125 \end_inset
2126
2127 Take care of device memory handling (allocation / freeing)
2128 \end_layout
2129
2130 \begin_layout Itemize
2131 \begin_inset ERT
2132 status collapsed
2133
2134 \begin_layout Plain Layout
2135
2136 <2->
2137 \end_layout
2138
2139 \end_inset
2140
2141 Copy data from host to device and from device back to host
2142 \end_layout
2143
2144 \begin_layout Itemize
2145 \begin_inset ERT
2146 status collapsed
2147
2148 \begin_layout Plain Layout
2149
2150 <3->
2151 \end_layout
2152
2153 \end_inset
2154
2155 Completely transparent to users of CUDA/Ada
2156 \end_layout
2157
2158 \begin_layout Itemize
2159 \begin_inset ERT
2160 status collapsed
2161
2162 \begin_layout Plain Layout
2163
2164 <4->
2165 \end_layout
2166
2167 \end_inset
2168
2169 Implemented using Ada generics
2170 \end_layout
2171
2172 \begin_layout Itemize
2173 \begin_inset ERT
2174 status collapsed
2175
2176 \begin_layout Plain Layout
2177
2178 <5->
2179 \end_layout
2180
2181 \end_inset
2182
2183 Three different argument types: 
2184 \end_layout
2185
2186 \begin_deeper
2187 \begin_layout Itemize
2188
2189 \shape italic
2190 In
2191 \end_layout
2192
2193 \begin_layout Itemize
2194
2195 \shape italic
2196 Out
2197 \end_layout
2198
2199 \begin_layout Itemize
2200
2201 \shape italic
2202 InOut
2203 \end_layout
2204
2205 \end_deeper
2206 \begin_layout Itemize
2207 \begin_inset ERT
2208 status collapsed
2209
2210 \begin_layout Plain Layout
2211
2212 <6->
2213 \end_layout
2214
2215 \end_inset
2216
2217 Similar to Ada's formal parameter modes (in, out, in out)
2218 \end_layout
2219
2220 \end_deeper
2221 \begin_layout EndFrame
2222
2223 \end_layout
2224
2225 \begin_layout Standard
2226 \begin_inset ERT
2227 status open
2228
2229 \begin_layout Plain Layout
2230
2231
2232 \backslash
2233 begin{frame}[fragile]
2234 \backslash
2235 frametitle{Kernel arguments II}
2236 \end_layout
2237
2238 \end_inset
2239
2240
2241 \end_layout
2242
2243 \begin_layout Standard
2244 \begin_inset listings
2245 inline false
2246 status open
2247
2248 \begin_layout Plain Layout
2249
2250 generic
2251 \end_layout
2252
2253 \begin_layout Plain Layout
2254
2255    type Data_Type (<>) is private;
2256 \end_layout
2257
2258 \begin_layout Plain Layout
2259
2260 package Arg_Creators is
2261 \end_layout
2262
2263 \begin_layout Plain Layout
2264
2265    function In_Arg (Data : Data_Type) return Arg_Type;
2266 \end_layout
2267
2268 \begin_layout Plain Layout
2269
2270 \end_layout
2271
2272 \begin_layout Plain Layout
2273
2274    function Out_Arg (Data : not null access Data_Type) return Arg_Type;
2275 \end_layout
2276
2277 \begin_layout Plain Layout
2278
2279 \end_layout
2280
2281 \begin_layout Plain Layout
2282
2283    function In_Out_Arg (Data : not null access Data_Type) return Arg_Type;
2284 \end_layout
2285
2286 \begin_layout Plain Layout
2287
2288 end Arg_Creators;
2289 \end_layout
2290
2291 \end_inset
2292
2293
2294 \end_layout
2295
2296 \begin_layout Standard
2297 \begin_inset ERT
2298 status open
2299
2300 \begin_layout Plain Layout
2301
2302
2303 \backslash
2304 end{frame}
2305 \end_layout
2306
2307 \end_inset
2308
2309
2310 \end_layout
2311
2312 \begin_layout Standard
2313 \begin_inset ERT
2314 status open
2315
2316 \begin_layout Plain Layout
2317
2318
2319 \backslash
2320 begin{frame}[fragile]
2321 \backslash
2322 frametitle{Kernel arguments III}
2323 \end_layout
2324
2325 \end_inset
2326
2327
2328 \end_layout
2329
2330 \begin_layout Standard
2331 \begin_inset listings
2332 inline false
2333 status open
2334
2335 \begin_layout Plain Layout
2336
2337    ...
2338 \end_layout
2339
2340 \begin_layout Plain Layout
2341
2342    package Matrix_Args is new CUDA.Compiler.Arg_Creators
2343 \end_layout
2344
2345 \begin_layout Plain Layout
2346
2347      (Data_Type => Ada.Numerics.Real_Arrays.Real_Matrix);
2348 \end_layout
2349
2350 \begin_layout Plain Layout
2351
2352    use Matrix_Args;
2353 \end_layout
2354
2355 \begin_layout Plain Layout
2356
2357 \end_layout
2358
2359 \begin_layout Plain Layout
2360
2361    Matrix : Ada.Numerics.Real_Arrays.Real_Matrix
2362 \end_layout
2363
2364 \begin_layout Plain Layout
2365
2366      := (1 ..
2367  N => (1 ..
2368  N => 0.0));
2369 \end_layout
2370
2371 \begin_layout Plain Layout
2372
2373    Arg    : CUDA.Compiler.Arg_Type
2374 \end_layout
2375
2376 \begin_layout Plain Layout
2377
2378      := In_Arg (Data => Matrix);
2379 \end_layout
2380
2381 \begin_layout Plain Layout
2382
2383 begin
2384 \end_layout
2385
2386 \begin_layout Plain Layout
2387
2388    ...
2389 \end_layout
2390
2391 \end_inset
2392
2393
2394 \end_layout
2395
2396 \begin_layout Standard
2397 \begin_inset ERT
2398 status open
2399
2400 \begin_layout Plain Layout
2401
2402
2403 \backslash
2404 end{frame}
2405 \end_layout
2406
2407 \end_inset
2408
2409
2410 \end_layout
2411
2412 \begin_layout Standard
2413 \begin_inset ERT
2414 status open
2415
2416 \begin_layout Plain Layout
2417
2418
2419 \backslash
2420 begin{frame}[fragile]
2421 \backslash
2422 frametitle{Kernel arguments IV}
2423 \end_layout
2424
2425 \end_inset
2426
2427
2428 \end_layout
2429
2430 \begin_layout Standard
2431
2432 \family typewriter
2433 \series bold
2434 Kernel signature
2435 \end_layout
2436
2437 \begin_layout Standard
2438 \begin_inset listings
2439 lstparams "language=C"
2440 inline false
2441 status open
2442
2443 \begin_layout Plain Layout
2444
2445 void mul(float* A, float* B, float* C)
2446 \end_layout
2447
2448 \end_inset
2449
2450
2451 \end_layout
2452
2453 \begin_layout Standard
2454
2455 \family typewriter
2456 \series bold
2457 CUDA/Ada call
2458 \end_layout
2459
2460 \begin_layout Standard
2461 \begin_inset listings
2462 inline false
2463 status open
2464
2465 \begin_layout Plain Layout
2466
2467    Func.Call 
2468 \end_layout
2469
2470 \begin_layout Plain Layout
2471
2472      (Args =>
2473 \end_layout
2474
2475 \begin_layout Plain Layout
2476
2477         (1 => In_Arg  (Data => A),
2478 \end_layout
2479
2480 \begin_layout Plain Layout
2481
2482          2 => In_Arg  (Data => B),
2483 \end_layout
2484
2485 \begin_layout Plain Layout
2486
2487          3 => Out_Arg (Data => C'Access)));
2488 \end_layout
2489
2490 \end_inset
2491
2492
2493 \end_layout
2494
2495 \begin_layout Standard
2496 \begin_inset ERT
2497 status open
2498
2499 \begin_layout Plain Layout
2500
2501
2502 \backslash
2503 end{frame}
2504 \end_layout
2505
2506 \end_inset
2507
2508
2509 \end_layout
2510
2511 \begin_layout Standard
2512 \begin_inset ERT
2513 status open
2514
2515 \begin_layout Plain Layout
2516
2517
2518 \backslash
2519 begin{frame}[fragile]
2520 \backslash
2521 frametitle{Error handling}
2522 \end_layout
2523
2524 \end_inset
2525
2526
2527 \end_layout
2528
2529 \begin_layout Block
2530 \begin_inset ERT
2531 status collapsed
2532
2533 \begin_layout Plain Layout
2534
2535 {
2536 \end_layout
2537
2538 \end_inset
2539
2540
2541 \family typewriter
2542 Functionality
2543 \family default
2544
2545 \begin_inset ERT
2546 status collapsed
2547
2548 \begin_layout Plain Layout
2549
2550 }
2551 \end_layout
2552
2553 \end_inset
2554
2555
2556 \end_layout
2557
2558 \begin_deeper
2559 \begin_layout Itemize
2560 Translation of CUDA errors to Ada exceptions
2561 \end_layout
2562
2563 \begin_layout Itemize
2564 Automated error checking of all low-level Thin-Binding calls
2565 \end_layout
2566
2567 \begin_layout Itemize
2568 Resolution of error code to error message 
2569 \end_layout
2570
2571 \end_deeper
2572 \begin_layout ExampleBlock
2573 \begin_inset ERT
2574 status collapsed
2575
2576 \begin_layout Plain Layout
2577
2578 {
2579 \end_layout
2580
2581 \end_inset
2582
2583 Requesting a nonexistent kernel 'Matrix_Mul'
2584 \begin_inset ERT
2585 status collapsed
2586
2587 \begin_layout Plain Layout
2588
2589 }
2590 \end_layout
2591
2592 \end_inset
2593
2594
2595 \end_layout
2596
2597 \begin_deeper
2598 \begin_layout Standard
2599 \begin_inset ERT
2600 status open
2601
2602 \begin_layout Plain Layout
2603
2604
2605 \backslash
2606 begin{verbatim}
2607 \end_layout
2608
2609 \begin_layout Plain Layout
2610
2611 Execution terminated by unhandled exception
2612 \end_layout
2613
2614 \begin_layout Plain Layout
2615
2616 Exception name: CUDA.CUDA_ERROR
2617 \end_layout
2618
2619 \begin_layout Plain Layout
2620
2621 Message: Could not get function Matrix_Mul (Not found)
2622 \end_layout
2623
2624 \begin_layout Plain Layout
2625
2626 Call stack traceback locations:
2627 \end_layout
2628
2629 \begin_layout Plain Layout
2630
2631 0x4079b6 0x40bc3a 0x406a4b 0x406299
2632 \end_layout
2633
2634 \begin_layout Plain Layout
2635
2636 0x7f159e4afc4b 0x405bd7
2637 \end_layout
2638
2639 \begin_layout Plain Layout
2640
2641
2642 \backslash
2643 end{verbatim}
2644 \end_layout
2645
2646 \end_inset
2647
2648
2649 \end_layout
2650
2651 \end_deeper
2652 \begin_layout Standard
2653 \begin_inset ERT
2654 status open
2655
2656 \begin_layout Plain Layout
2657
2658
2659 \backslash
2660 end{frame}
2661 \end_layout
2662
2663 \end_inset
2664
2665
2666 \end_layout
2667
2668 \begin_layout Standard
2669 \begin_inset ERT
2670 status open
2671
2672 \begin_layout Plain Layout
2673
2674
2675 \backslash
2676 begin{frame}[fragile]
2677 \backslash
2678 frametitle{Querying CUDA devices}
2679 \end_layout
2680
2681 \end_inset
2682
2683
2684 \end_layout
2685
2686 \begin_layout Block
2687 \begin_inset ERT
2688 status collapsed
2689
2690 \begin_layout Plain Layout
2691
2692 {
2693 \end_layout
2694
2695 \end_inset
2696
2697
2698 \family typewriter
2699 Functionality
2700 \family default
2701
2702 \begin_inset ERT
2703 status collapsed
2704
2705 \begin_layout Plain Layout
2706
2707 }
2708 \end_layout
2709
2710 \end_inset
2711
2712
2713 \end_layout
2714
2715 \begin_deeper
2716 \begin_layout Itemize
2717 Enumerate all available CUDA devices
2718 \end_layout
2719
2720 \begin_layout Itemize
2721 Query device properties like name, compute capability, etc.
2722 \end_layout
2723
2724 \end_deeper
2725 \begin_layout Separator
2726
2727 \end_layout
2728
2729 \begin_layout Standard
2730 \begin_inset listings
2731 inline false
2732 status open
2733
2734 \begin_layout Plain Layout
2735
2736 with Ada.Text_IO;
2737 \end_layout
2738
2739 \begin_layout Plain Layout
2740
2741 with CUDA.Driver; use CUDA.Driver;
2742 \end_layout
2743
2744 \begin_layout Plain Layout
2745
2746 \end_layout
2747
2748 \begin_layout Plain Layout
2749
2750 procedure Enum_Devices is
2751 \end_layout
2752
2753 \begin_layout Plain Layout
2754
2755    procedure Print_Name (Dev : Device_Type) is
2756 \end_layout
2757
2758 \begin_layout Plain Layout
2759
2760    begin
2761 \end_layout
2762
2763 \begin_layout Plain Layout
2764
2765       Ada.Text_IO.Put_Line (Name (Dev));
2766 \end_layout
2767
2768 \begin_layout Plain Layout
2769
2770    end Print_Name;
2771 \end_layout
2772
2773 \begin_layout Plain Layout
2774
2775 begin
2776 \end_layout
2777
2778 \begin_layout Plain Layout
2779
2780    Iterate (Process => Print_Name'Access);
2781 \end_layout
2782
2783 \begin_layout Plain Layout
2784
2785 end Enum_Devices;
2786 \end_layout
2787
2788 \end_inset
2789
2790
2791 \end_layout
2792
2793 \begin_layout Standard
2794 \begin_inset ERT
2795 status open
2796
2797 \begin_layout Plain Layout
2798
2799
2800 \backslash
2801 end{frame}
2802 \end_layout
2803
2804 \end_inset
2805
2806
2807 \end_layout
2808
2809 \begin_layout Section
2810 Conclusion
2811 \end_layout
2812
2813 \begin_layout Subsection
2814 Performance
2815 \end_layout
2816
2817 \begin_layout BeginFrame
2818 Performance
2819 \end_layout
2820
2821 \begin_layout Block
2822 \begin_inset ERT
2823 status collapsed
2824
2825 \begin_layout Plain Layout
2826
2827 {
2828 \end_layout
2829
2830 \end_inset
2831
2832
2833 \family typewriter
2834 Ambition
2835 \family default
2836
2837 \begin_inset ERT
2838 status collapsed
2839
2840 \begin_layout Plain Layout
2841
2842 }
2843 \end_layout
2844
2845 \end_inset
2846
2847
2848 \end_layout
2849
2850 \begin_deeper
2851 \begin_layout Quote
2852 Programs using CUDA/Ada should reach 
2853 \begin_inset Quotes eld
2854 \end_inset
2855
2856 native
2857 \begin_inset Quotes erd
2858 \end_inset
2859
2860  CUDA speed.
2861  The overhead imposed by Ada should be kept minimal.
2862 \end_layout
2863
2864 \end_deeper
2865 \begin_layout Separator
2866
2867 \end_layout
2868
2869 \begin_layout Block
2870 \begin_inset ERT
2871 status collapsed
2872
2873 \begin_layout Plain Layout
2874
2875 {
2876 \end_layout
2877
2878 \end_inset
2879
2880
2881 \family typewriter
2882 Measurement
2883 \family default
2884  methodology
2885 \begin_inset ERT
2886 status collapsed
2887
2888 \begin_layout Plain Layout
2889
2890 }
2891 \end_layout
2892
2893 \end_inset
2894
2895
2896 \end_layout
2897
2898 \begin_deeper
2899 \begin_layout Itemize
2900 \begin_inset ERT
2901 status collapsed
2902
2903 \begin_layout Plain Layout
2904
2905 <1->
2906 \end_layout
2907
2908 \end_inset
2909
2910 Cumulated runtime of 20 matrix multiplications (512 x 512)
2911 \end_layout
2912
2913 \begin_layout Itemize
2914 \begin_inset ERT
2915 status collapsed
2916
2917 \begin_layout Plain Layout
2918
2919 <2->
2920 \end_layout
2921
2922 \end_inset
2923
2924 Comparison of different implementations:
2925 \end_layout
2926
2927 \begin_deeper
2928 \begin_layout Itemize
2929 Ada (CPU)
2930 \end_layout
2931
2932 \begin_layout Itemize
2933 CUDA Runtime API
2934 \end_layout
2935
2936 \begin_layout Itemize
2937 CUDA Driver API
2938 \end_layout
2939
2940 \begin_layout Itemize
2941 CUDA/Ada
2942 \end_layout
2943
2944 \end_deeper
2945 \begin_layout Itemize
2946 \begin_inset ERT
2947 status collapsed
2948
2949 \begin_layout Plain Layout
2950
2951 <3->
2952 \end_layout
2953
2954 \end_inset
2955
2956 Same grid and block size for all CUDA implementations
2957 \end_layout
2958
2959 \end_deeper
2960 \begin_layout BeginFrame
2961 Performance II
2962 \end_layout
2963
2964 \begin_layout Standard
2965 \begin_inset Float table
2966 wide false
2967 sideways false
2968 status open
2969
2970 \begin_layout Plain Layout
2971 \align center
2972 \begin_inset Tabular
2973 <lyxtabular version="3" rows="7" columns="2">
2974 <features tabularvalignment="middle">
2975 <column alignment="center" valignment="top" width="0">
2976 <column alignment="center" valignment="top" width="0pt">
2977 <row>
2978 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
2979 \begin_inset Text
2980
2981 \begin_layout Plain Layout
2982 Processor
2983 \end_layout
2984
2985 \end_inset
2986 </cell>
2987 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
2988 \begin_inset Text
2989
2990 \begin_layout Plain Layout
2991 AMD Phenom II X4 940
2992 \end_layout
2993
2994 \end_inset
2995 </cell>
2996 </row>
2997 <row>
2998 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
2999 \begin_inset Text
3000
3001 \begin_layout Plain Layout
3002 Graphics Card
3003 \end_layout
3004
3005 \end_inset
3006 </cell>
3007 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3008 \begin_inset Text
3009
3010 \begin_layout Plain Layout
3011 GeForce GTX 560 Ti
3012 \end_layout
3013
3014 \end_inset
3015 </cell>
3016 </row>
3017 <row>
3018 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3019 \begin_inset Text
3020
3021 \begin_layout Plain Layout
3022 Operating System
3023 \end_layout
3024
3025 \end_inset
3026 </cell>
3027 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3028 \begin_inset Text
3029
3030 \begin_layout Plain Layout
3031 Debian Linux 6.0
3032 \end_layout
3033
3034 \end_inset
3035 </cell>
3036 </row>
3037 <row>
3038 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3039 \begin_inset Text
3040
3041 \begin_layout Plain Layout
3042 Kernel
3043 \end_layout
3044
3045 \end_inset
3046 </cell>
3047 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3048 \begin_inset Text
3049
3050 \begin_layout Plain Layout
3051  2.6.32-5-amd64
3052 \end_layout
3053
3054 \end_inset
3055 </cell>
3056 </row>
3057 <row>
3058 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3059 \begin_inset Text
3060
3061 \begin_layout Plain Layout
3062 Ada Compiler
3063 \end_layout
3064
3065 \end_inset
3066 </cell>
3067 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3068 \begin_inset Text
3069
3070 \begin_layout Plain Layout
3071 FSF GNAT 4.4.5
3072 \end_layout
3073
3074 \end_inset
3075 </cell>
3076 </row>
3077 <row>
3078 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3079 \begin_inset Text
3080
3081 \begin_layout Plain Layout
3082 NVIDIA Graphics Driver
3083 \end_layout
3084
3085 \end_inset
3086 </cell>
3087 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3088 \begin_inset Text
3089
3090 \begin_layout Plain Layout
3091 270.41.19, Linux 64-bit
3092 \end_layout
3093
3094 \end_inset
3095 </cell>
3096 </row>
3097 <row>
3098 <cell alignment="center" valignment="top" topline="true" bottomline="true" leftline="true" usebox="none">
3099 \begin_inset Text
3100
3101 \begin_layout Plain Layout
3102 NVIDIA CUDA Toolkit
3103 \end_layout
3104
3105 \end_inset
3106 </cell>
3107 <cell alignment="center" valignment="top" topline="true" bottomline="true" leftline="true" rightline="true" usebox="none">
3108 \begin_inset Text
3109
3110 \begin_layout Plain Layout
3111 4.0.17, Linux 64-bit
3112 \end_layout
3113
3114 \end_inset
3115 </cell>
3116 </row>
3117 </lyxtabular>
3118
3119 \end_inset
3120
3121
3122 \begin_inset Caption
3123
3124 \begin_layout Plain Layout
3125 Test system specs
3126 \end_layout
3127
3128 \end_inset
3129
3130
3131 \end_layout
3132
3133 \end_inset
3134
3135
3136 \end_layout
3137
3138 \begin_layout EndFrame
3139
3140 \end_layout
3141
3142 \begin_layout BeginFrame
3143 Benchmarking results
3144 \end_layout
3145
3146 \begin_layout Standard
3147 \align center
3148 \begin_inset Graphics
3149         filename performance-chart.pdf
3150         scale 76
3151
3152 \end_inset
3153
3154
3155 \end_layout
3156
3157 \begin_layout EndFrame
3158
3159 \end_layout
3160
3161 \begin_layout BeginFrame
3162 Performance III
3163 \end_layout
3164
3165 \begin_layout Block
3166 \begin_inset ERT
3167 status collapsed
3168
3169 \begin_layout Plain Layout
3170
3171 {
3172 \end_layout
3173
3174 \end_inset
3175
3176
3177 \family typewriter
3178 Interpretation
3179 \family default
3180
3181 \begin_inset ERT
3182 status collapsed
3183
3184 \begin_layout Plain Layout
3185
3186 }
3187 \end_layout
3188
3189 \end_inset
3190
3191
3192 \end_layout
3193
3194 \begin_deeper
3195 \begin_layout Itemize
3196 \begin_inset ERT
3197 status collapsed
3198
3199 \begin_layout Plain Layout
3200
3201 <1->
3202 \end_layout
3203
3204 \end_inset
3205
3206 Native Ada (CPU) is slow: CPU vs.
3207  GPU/CUDA
3208 \end_layout
3209
3210 \begin_layout Itemize
3211 \begin_inset ERT
3212 status collapsed
3213
3214 \begin_layout Plain Layout
3215
3216 <2->
3217 \end_layout
3218
3219 \end_inset
3220
3221 CUDA/Ada is faster than CUDA Runtime API:
3222 \end_layout
3223
3224 \begin_deeper
3225 \begin_layout Itemize
3226 CUDA Runtime API generates management as well as kernel launch code, etc.
3227 \end_layout
3228
3229 \begin_layout Itemize
3230 CUDA/Ada performs bare minimum of management operations
3231 \end_layout
3232
3233 \end_deeper
3234 \begin_layout Itemize
3235 \begin_inset ERT
3236 status collapsed
3237
3238 \begin_layout Plain Layout
3239
3240 <3->
3241 \end_layout
3242
3243 \end_inset
3244
3245 CUDA/Ada is negligibly slower than CUDA Driver API
3246 \end_layout
3247
3248 \begin_layout Itemize
3249 \begin_inset ERT
3250 status collapsed
3251
3252 \begin_layout Plain Layout
3253
3254 <4->
3255 \end_layout
3256
3257 \end_inset
3258
3259 No visible performance penalty
3260 \end_layout
3261
3262 \end_deeper
3263 \begin_layout EndFrame
3264
3265 \end_layout
3266
3267 \begin_layout Subsection
3268 Review
3269 \end_layout
3270
3271 \begin_layout BeginFrame
3272 Review
3273 \end_layout
3274
3275 \begin_layout Block
3276 \begin_inset ERT
3277 status collapsed
3278
3279 \begin_layout Plain Layout
3280
3281 {
3282 \end_layout
3283
3284 \end_inset
3285
3286 Assessment of results
3287 \begin_inset ERT
3288 status collapsed
3289
3290 \begin_layout Plain Layout
3291
3292 }
3293 \end_layout
3294
3295 \end_inset
3296
3297
3298 \end_layout
3299
3300 \begin_deeper
3301 \begin_layout Itemize
3302 \begin_inset ERT
3303 status collapsed
3304
3305 \begin_layout Plain Layout
3306
3307 <1->
3308 \end_layout
3309
3310 \end_inset
3311
3312 Thick-Binding provides easy usage of CUDA from Ada
3313 \end_layout
3314
3315 \begin_layout Itemize
3316 \begin_inset ERT
3317 status collapsed
3318
3319 \begin_layout Plain Layout
3320
3321 <2->
3322 \end_layout
3323
3324 \end_inset
3325
3326 High level abstractions go well with other Ada language constructs
3327 \end_layout
3328
3329 \begin_layout Itemize
3330 \begin_inset ERT
3331 status collapsed
3332
3333 \begin_layout Plain Layout
3334
3335 <3->
3336 \end_layout
3337
3338 \end_inset
3339
3340 Simple Autoinitialization of CUDA via 
3341 \family typewriter
3342 Autoinit
3343 \family default
3344  package
3345 \end_layout
3346
3347 \begin_layout Itemize
3348 \begin_inset ERT
3349 status collapsed
3350
3351 \begin_layout Plain Layout
3352
3353 <4->
3354 \end_layout
3355
3356 \end_inset
3357
3358 JIT compilation of kernels provides flexibilty and agility
3359 \end_layout
3360
3361 \begin_layout Itemize
3362 \begin_inset ERT
3363 status collapsed
3364
3365 \begin_layout Plain Layout
3366
3367 <5->
3368 \end_layout
3369
3370 \end_inset
3371
3372 Speed of CUDA/Ada is excellent
3373 \end_layout
3374
3375 \begin_layout Itemize
3376 \begin_inset ERT
3377 status collapsed
3378
3379 \begin_layout Plain Layout
3380
3381 <6->
3382 \end_layout
3383
3384 \end_inset
3385
3386 Automated Thin-Binding generation allows for easy integration of future
3387  CUDA API changes
3388 \end_layout
3389
3390 \begin_layout Itemize
3391 \begin_inset ERT
3392 status collapsed
3393
3394 \begin_layout Plain Layout
3395
3396 <7->
3397 \end_layout
3398
3399 \end_inset
3400
3401 Paper documents how to write a modern, maintainable language binding for
3402  Ada in general
3403 \end_layout
3404
3405 \end_deeper
3406 \begin_layout EndFrame
3407
3408 \end_layout
3409
3410 \begin_layout BeginFrame
3411 Review II
3412 \end_layout
3413
3414 \begin_layout Block
3415 \begin_inset ERT
3416 status collapsed
3417
3418 \begin_layout Plain Layout
3419
3420 {
3421 \end_layout
3422
3423 \end_inset
3424
3425 Deliverables
3426 \begin_inset ERT
3427 status collapsed
3428
3429 \begin_layout Plain Layout
3430
3431 }
3432 \end_layout
3433
3434 \end_inset
3435
3436
3437 \end_layout
3438
3439 \begin_deeper
3440 \begin_layout Itemize
3441 Paper
3442 \end_layout
3443
3444 \begin_layout Itemize
3445 Slides
3446 \end_layout
3447
3448 \begin_layout Itemize
3449 Source code (GPLv3+)
3450 \family typewriter
3451
3452 \begin_inset Newline newline
3453 \end_inset
3454
3455
3456 \family default
3457
3458 \begin_inset Flex URL
3459 status collapsed
3460
3461 \begin_layout Plain Layout
3462
3463 http://git.codelabs.ch/?p=cuda-ada.git
3464 \end_layout
3465
3466 \end_inset
3467
3468
3469 \family typewriter
3470
3471 \begin_inset Newline newline
3472 \end_inset
3473
3474 git clone http://git.codelabs.ch/git/cuda-ada.git
3475 \end_layout
3476
3477 \begin_layout Itemize
3478 Website with all documents and additional information: 
3479 \begin_inset Flex URL
3480 status collapsed
3481
3482 \begin_layout Plain Layout
3483
3484 http://www.codelabs.ch/cuda-ada/
3485 \end_layout
3486
3487 \end_inset
3488
3489
3490 \end_layout
3491
3492 \end_deeper
3493 \begin_layout EndFrame
3494
3495 \end_layout
3496
3497 \begin_layout Subsection
3498 Outlook
3499 \end_layout
3500
3501 \begin_layout BeginFrame
3502 Outlook
3503 \end_layout
3504
3505 \begin_layout Block
3506 \begin_inset ERT
3507 status collapsed
3508
3509 \begin_layout Plain Layout
3510
3511 {
3512 \end_layout
3513
3514 \end_inset
3515
3516 CUDA/Ada
3517 \begin_inset ERT
3518 status collapsed
3519
3520 \begin_layout Plain Layout
3521
3522 }
3523 \end_layout
3524
3525 \end_inset
3526
3527
3528 \end_layout
3529
3530 \begin_deeper
3531 \begin_layout Itemize
3532 Implement additional features:
3533 \end_layout
3534
3535 \begin_deeper
3536 \begin_layout Itemize
3537 Abstractions for kernel generation (e.g.
3538  element-wise)
3539 \end_layout
3540
3541 \begin_layout Itemize
3542 Kernel grid- and block size depending on device capabilities
3543 \end_layout
3544
3545 \begin_layout Itemize
3546 Multi-Device support
3547 \end_layout
3548
3549 \begin_layout Itemize
3550 ...
3551 \end_layout
3552
3553 \end_deeper
3554 \begin_layout Itemize
3555 Announce project to Ada community
3556 \end_layout
3557
3558 \end_deeper
3559 \begin_layout Separator
3560
3561 \end_layout
3562
3563 \begin_layout Block
3564 \begin_inset ERT
3565 status collapsed
3566
3567 \begin_layout Plain Layout
3568
3569 {
3570 \end_layout
3571
3572 \end_inset
3573
3574 Ada and CUDA
3575 \begin_inset ERT
3576 status collapsed
3577
3578 \begin_layout Plain Layout
3579
3580 }
3581 \end_layout
3582
3583 \end_inset
3584
3585
3586 \end_layout
3587
3588 \begin_deeper
3589 \begin_layout Itemize
3590 Study NVIDIA's LLVM-based CUDA compiler (announced, not yet released)
3591 \end_layout
3592
3593 \end_deeper
3594 \begin_layout EndFrame
3595
3596 \end_layout
3597
3598 \begin_layout BeginFrame
3599 Questions
3600 \end_layout
3601
3602 \begin_layout Standard
3603 \align center
3604
3605 \size giant
3606 Thank you for your attention!
3607 \end_layout
3608
3609 \begin_layout EndFrame
3610
3611 \end_layout
3612
3613 \end_body
3614 \end_document