Doc: Minor corrections to slides
[cuda-ada.git] / doc / slides.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: Program 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 effort 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 \family typewriter
1409 unsigned_long_long
1410 \end_layout
1411
1412 \end_deeper
1413 \end_deeper
1414 \begin_layout Standard
1415 \begin_inset ERT
1416 status open
1417
1418 \begin_layout Plain Layout
1419
1420
1421 \backslash
1422 begin{verbatim}
1423 \end_layout
1424
1425 \begin_layout Plain Layout
1426
1427 gcc -c -fdump-ada-spec cuda.h
1428 \end_layout
1429
1430 \begin_layout Plain Layout
1431
1432 gcc -c -fdump-ada-spec cuda_runtime.h
1433 \end_layout
1434
1435 \begin_layout Plain Layout
1436
1437
1438 \backslash
1439 end{verbatim}
1440 \end_layout
1441
1442 \end_inset
1443
1444
1445 \end_layout
1446
1447 \begin_layout Standard
1448 \begin_inset ERT
1449 status open
1450
1451 \begin_layout Plain Layout
1452
1453
1454 \backslash
1455 end{frame}
1456 \end_layout
1457
1458 \end_inset
1459
1460
1461 \end_layout
1462
1463 \begin_layout Standard
1464 \begin_inset ERT
1465 status open
1466
1467 \begin_layout Plain Layout
1468
1469
1470 \backslash
1471 begin{frame}[fragile]
1472 \backslash
1473 frametitle{Architecture (build logic)}
1474 \end_layout
1475
1476 \end_inset
1477
1478
1479 \end_layout
1480
1481 \begin_layout Standard
1482 \begin_inset ERT
1483 status open
1484
1485 \begin_layout Plain Layout
1486
1487
1488 \backslash
1489 begin{verbatim}
1490 \end_layout
1491
1492 \begin_layout Plain Layout
1493
1494 ARCH ?= $(shell uname -m)
1495 \end_layout
1496
1497 \begin_layout Plain Layout
1498
1499 gnatmake -Pcuda -XARCH=$(ARCH)
1500 \end_layout
1501
1502 \begin_layout Plain Layout
1503
1504
1505 \backslash
1506 end{verbatim}
1507 \end_layout
1508
1509 \end_inset
1510
1511
1512 \end_layout
1513
1514 \begin_layout Standard
1515 \begin_inset listings
1516 inline false
1517 status open
1518
1519 \begin_layout Plain Layout
1520
1521 with "thin/binding";
1522 \end_layout
1523
1524 \end_inset
1525
1526
1527 \end_layout
1528
1529 \begin_layout Standard
1530 \begin_inset listings
1531 inline false
1532 status open
1533
1534 \begin_layout Plain Layout
1535
1536 type Arch_Type is ("x86_64", "i686"); 
1537 \end_layout
1538
1539 \begin_layout Plain Layout
1540
1541 Arch : Arch_Type := external ("ARCH", "x86_64");
1542 \end_layout
1543
1544 \begin_layout Plain Layout
1545
1546 \end_layout
1547
1548 \begin_layout Plain Layout
1549
1550 for Source_Dirs use (".", ARCH);
1551 \end_layout
1552
1553 \end_inset
1554
1555
1556 \end_layout
1557
1558 \begin_layout Standard
1559 \begin_inset ERT
1560 status open
1561
1562 \begin_layout Plain Layout
1563
1564
1565 \backslash
1566 end{frame}
1567 \end_layout
1568
1569 \end_inset
1570
1571
1572 \end_layout
1573
1574 \begin_layout Subsection
1575 Thick-Binding
1576 \end_layout
1577
1578 \begin_layout Standard
1579 \begin_inset ERT
1580 status open
1581
1582 \begin_layout Plain Layout
1583
1584
1585 \backslash
1586 begin{frame}[fragile]
1587 \backslash
1588 frametitle{Autoinit}
1589 \end_layout
1590
1591 \end_inset
1592
1593
1594 \end_layout
1595
1596 \begin_layout Block
1597 \begin_inset ERT
1598 status collapsed
1599
1600 \begin_layout Plain Layout
1601
1602 {
1603 \end_layout
1604
1605 \end_inset
1606
1607
1608 \family typewriter
1609 Functionality
1610 \family default
1611
1612 \begin_inset ERT
1613 status collapsed
1614
1615 \begin_layout Plain Layout
1616
1617 }
1618 \end_layout
1619
1620 \end_inset
1621
1622
1623 \end_layout
1624
1625 \begin_deeper
1626 \begin_layout Itemize
1627 Takes care of CUDA initialisation task
1628 \end_layout
1629
1630 \begin_deeper
1631 \begin_layout Itemize
1632 cuInit
1633 \end_layout
1634
1635 \begin_layout Itemize
1636 cuDeviceGet
1637 \end_layout
1638
1639 \begin_layout Itemize
1640 cuCtxCreate
1641 \end_layout
1642
1643 \end_deeper
1644 \begin_layout Itemize
1645 Handles release of CUDA context
1646 \end_layout
1647
1648 \begin_deeper
1649 \begin_layout Itemize
1650 cuCtxDestroy
1651 \end_layout
1652
1653 \end_deeper
1654 \end_deeper
1655 \begin_layout Standard
1656 \begin_inset listings
1657 inline false
1658 status open
1659
1660 \begin_layout Plain Layout
1661
1662 with CUDA.Autoinit;
1663 \end_layout
1664
1665 \end_inset
1666
1667
1668 \end_layout
1669
1670 \begin_layout Standard
1671 \begin_inset ERT
1672 status open
1673
1674 \begin_layout Plain Layout
1675
1676
1677 \backslash
1678 end{frame}
1679 \end_layout
1680
1681 \end_inset
1682
1683
1684 \end_layout
1685
1686 \begin_layout Standard
1687 \begin_inset ERT
1688 status open
1689
1690 \begin_layout Plain Layout
1691
1692
1693 \backslash
1694 begin{frame}[fragile]
1695 \backslash
1696 frametitle{Source-Modules}
1697 \end_layout
1698
1699 \end_inset
1700
1701
1702 \end_layout
1703
1704 \begin_layout Block
1705 \begin_inset ERT
1706 status collapsed
1707
1708 \begin_layout Plain Layout
1709
1710 {
1711 \end_layout
1712
1713 \end_inset
1714
1715
1716 \family typewriter
1717 Functionality
1718 \family default
1719
1720 \begin_inset ERT
1721 status collapsed
1722
1723 \begin_layout Plain Layout
1724
1725 }
1726 \end_layout
1727
1728 \end_inset
1729
1730
1731 \end_layout
1732
1733 \begin_deeper
1734 \begin_layout Itemize
1735 Used to define CUDA kernels 
1736 \begin_inset Quotes eld
1737 \end_inset
1738
1739 inline
1740 \begin_inset Quotes erd
1741 \end_inset
1742
1743
1744 \end_layout
1745
1746 \end_deeper
1747 \begin_layout Standard
1748 \begin_inset listings
1749 inline false
1750 status open
1751
1752 \begin_layout Plain Layout
1753
1754 N   : constant                    := 32 * 1024;
1755 \end_layout
1756
1757 \begin_layout Plain Layout
1758
1759 Src : Compiler.Source_Module_Type :=
1760 \end_layout
1761
1762 \begin_layout Plain Layout
1763
1764  Compiler.Create
1765 \end_layout
1766
1767 \begin_layout Plain Layout
1768
1769   (Preamble  => "#define N" & N'Img,
1770 \end_layout
1771
1772 \begin_layout Plain Layout
1773
1774    Operation => "__global__ void add(float *a, float *b) {" 
1775 \end_layout
1776
1777 \begin_layout Plain Layout
1778
1779     & "  int tid = blockIdx.x;"                         
1780 \end_layout
1781
1782 \begin_layout Plain Layout
1783
1784     & "  while (tid < N) {"
1785 \end_layout
1786
1787 \begin_layout Plain Layout
1788
1789     & "    b[tid] = a[tid] + 10;"                       
1790 \end_layout
1791
1792 \begin_layout Plain Layout
1793
1794     & "    tid += gridDim.x;"
1795 \end_layout
1796
1797 \begin_layout Plain Layout
1798
1799     & "}}");
1800 \end_layout
1801
1802 \end_inset
1803
1804
1805 \end_layout
1806
1807 \begin_layout Standard
1808 \begin_inset ERT
1809 status open
1810
1811 \begin_layout Plain Layout
1812
1813
1814 \backslash
1815 end{frame}
1816 \end_layout
1817
1818 \end_inset
1819
1820
1821 \end_layout
1822
1823 \begin_layout Standard
1824 \begin_inset ERT
1825 status open
1826
1827 \begin_layout Plain Layout
1828
1829
1830 \backslash
1831 begin{frame}[fragile]
1832 \backslash
1833 frametitle{JIT-Compiler}
1834 \end_layout
1835
1836 \end_inset
1837
1838
1839 \end_layout
1840
1841 \begin_layout Block
1842 \begin_inset ERT
1843 status collapsed
1844
1845 \begin_layout Plain Layout
1846
1847 {
1848 \end_layout
1849
1850 \end_inset
1851
1852
1853 \family typewriter
1854 Functionality
1855 \family default
1856
1857 \begin_inset ERT
1858 status collapsed
1859
1860 \begin_layout Plain Layout
1861
1862 }
1863 \end_layout
1864
1865 \end_inset
1866
1867
1868 \end_layout
1869
1870 \begin_deeper
1871 \begin_layout Itemize
1872 Compile CUDA kernels at runtime
1873 \end_layout
1874
1875 \begin_layout Itemize
1876 Uses 
1877 \family typewriter
1878 nvcc
1879 \family default
1880  to generate CUBIN binary code
1881 \end_layout
1882
1883 \begin_layout Itemize
1884 Upload modules to GPU
1885 \end_layout
1886
1887 \begin_layout Itemize
1888 Caching of compiled modules
1889 \end_layout
1890
1891 \end_deeper
1892 \begin_layout Standard
1893 \begin_inset listings
1894 inline false
1895 status open
1896
1897 \begin_layout Plain Layout
1898
1899    ...
1900 \end_layout
1901
1902 \begin_layout Plain Layout
1903
1904 is
1905 \end_layout
1906
1907 \begin_layout Plain Layout
1908
1909    Module : Compiler.Module_Type; 
1910 \end_layout
1911
1912 \begin_layout Plain Layout
1913
1914 begin
1915 \end_layout
1916
1917 \begin_layout Plain Layout
1918
1919    Module := Compiler.Compile (Source => Src); 
1920 \end_layout
1921
1922 \begin_layout Plain Layout
1923
1924    ...
1925  
1926 \end_layout
1927
1928 \end_inset
1929
1930
1931 \end_layout
1932
1933 \begin_layout Standard
1934 \begin_inset ERT
1935 status open
1936
1937 \begin_layout Plain Layout
1938
1939
1940 \backslash
1941 end{frame}
1942 \end_layout
1943
1944 \end_inset
1945
1946
1947 \end_layout
1948
1949 \begin_layout BeginFrame
1950 JIT Workflow
1951 \end_layout
1952
1953 \begin_layout Standard
1954 \align center
1955 \begin_inset Graphics
1956         filename jit-compiler.eps
1957         scale 53
1958
1959 \end_inset
1960
1961
1962 \end_layout
1963
1964 \begin_layout EndFrame
1965
1966 \end_layout
1967
1968 \begin_layout Standard
1969 \begin_inset ERT
1970 status open
1971
1972 \begin_layout Plain Layout
1973
1974
1975 \backslash
1976 begin{frame}[fragile]
1977 \backslash
1978 frametitle{Calling a function}
1979 \end_layout
1980
1981 \end_inset
1982
1983
1984 \end_layout
1985
1986 \begin_layout Standard
1987 \begin_inset listings
1988 inline false
1989 status open
1990
1991 \begin_layout Plain Layout
1992
1993    ...
1994 \end_layout
1995
1996 \begin_layout Plain Layout
1997
1998    Func   : Compiler.Function_Type;
1999 \end_layout
2000
2001 \begin_layout Plain Layout
2002
2003    Module : Compiler.Module_Type; 
2004 \end_layout
2005
2006 \begin_layout Plain Layout
2007
2008 begin
2009 \end_layout
2010
2011 \begin_layout Plain Layout
2012
2013    Module := Compiler.Compile (Source => Src); 
2014 \end_layout
2015
2016 \begin_layout Plain Layout
2017
2018    Func   := Compiler.Get_Function
2019 \end_layout
2020
2021 \begin_layout Plain Layout
2022
2023      (Module => Module,
2024 \end_layout
2025
2026 \begin_layout Plain Layout
2027
2028       Name   => "add");
2029 \end_layout
2030
2031 \begin_layout Plain Layout
2032
2033 \end_layout
2034
2035 \begin_layout Plain Layout
2036
2037    Func.Call 
2038 \end_layout
2039
2040 \begin_layout Plain Layout
2041
2042      (Args =>
2043 \end_layout
2044
2045 \begin_layout Plain Layout
2046
2047         (1 => In_Arg (Data => A),
2048 \end_layout
2049
2050 \begin_layout Plain Layout
2051
2052          2 => In_Arg (Data => B),
2053 \end_layout
2054
2055 \begin_layout Plain Layout
2056
2057          3 => Out_Arg (Data => C'Access)));
2058 \end_layout
2059
2060 \end_inset
2061
2062
2063 \end_layout
2064
2065 \begin_layout Standard
2066 \begin_inset ERT
2067 status open
2068
2069 \begin_layout Plain Layout
2070
2071
2072 \backslash
2073 end{frame}
2074 \end_layout
2075
2076 \end_inset
2077
2078
2079 \end_layout
2080
2081 \begin_layout BeginFrame
2082 Kernel arguments
2083 \end_layout
2084
2085 \begin_layout Block
2086 \begin_inset ERT
2087 status collapsed
2088
2089 \begin_layout Plain Layout
2090
2091 {
2092 \end_layout
2093
2094 \end_inset
2095
2096
2097 \family typewriter
2098 Functionality
2099 \family default
2100
2101 \begin_inset ERT
2102 status collapsed
2103
2104 \begin_layout Plain Layout
2105
2106 }
2107 \end_layout
2108
2109 \end_inset
2110
2111
2112 \end_layout
2113
2114 \begin_deeper
2115 \begin_layout Itemize
2116 \begin_inset ERT
2117 status collapsed
2118
2119 \begin_layout Plain Layout
2120
2121 <1->
2122 \end_layout
2123
2124 \end_inset
2125
2126 Take care of device memory handling (allocation / freeing)
2127 \end_layout
2128
2129 \begin_layout Itemize
2130 \begin_inset ERT
2131 status collapsed
2132
2133 \begin_layout Plain Layout
2134
2135 <2->
2136 \end_layout
2137
2138 \end_inset
2139
2140 Copy data from host to device and from device back to host
2141 \end_layout
2142
2143 \begin_layout Itemize
2144 \begin_inset ERT
2145 status collapsed
2146
2147 \begin_layout Plain Layout
2148
2149 <3->
2150 \end_layout
2151
2152 \end_inset
2153
2154 Completely transparent to users of CUDA/Ada
2155 \end_layout
2156
2157 \begin_layout Itemize
2158 \begin_inset ERT
2159 status collapsed
2160
2161 \begin_layout Plain Layout
2162
2163 <4->
2164 \end_layout
2165
2166 \end_inset
2167
2168 Implemented using Ada generics
2169 \end_layout
2170
2171 \begin_layout Itemize
2172 \begin_inset ERT
2173 status collapsed
2174
2175 \begin_layout Plain Layout
2176
2177 <5->
2178 \end_layout
2179
2180 \end_inset
2181
2182 Three different argument types: 
2183 \end_layout
2184
2185 \begin_deeper
2186 \begin_layout Itemize
2187
2188 \shape italic
2189 In
2190 \end_layout
2191
2192 \begin_layout Itemize
2193
2194 \shape italic
2195 Out
2196 \end_layout
2197
2198 \begin_layout Itemize
2199
2200 \shape italic
2201 InOut
2202 \end_layout
2203
2204 \end_deeper
2205 \begin_layout Itemize
2206 \begin_inset ERT
2207 status collapsed
2208
2209 \begin_layout Plain Layout
2210
2211 <6->
2212 \end_layout
2213
2214 \end_inset
2215
2216 Similar to Ada's formal parameter modes (in, out, in out)
2217 \end_layout
2218
2219 \end_deeper
2220 \begin_layout EndFrame
2221
2222 \end_layout
2223
2224 \begin_layout Standard
2225 \begin_inset ERT
2226 status open
2227
2228 \begin_layout Plain Layout
2229
2230
2231 \backslash
2232 begin{frame}[fragile]
2233 \backslash
2234 frametitle{Kernel arguments II}
2235 \end_layout
2236
2237 \end_inset
2238
2239
2240 \end_layout
2241
2242 \begin_layout Standard
2243 \begin_inset listings
2244 inline false
2245 status open
2246
2247 \begin_layout Plain Layout
2248
2249 generic
2250 \end_layout
2251
2252 \begin_layout Plain Layout
2253
2254    type Data_Type is private;
2255 \end_layout
2256
2257 \begin_layout Plain Layout
2258
2259 package Arg_Creators is
2260 \end_layout
2261
2262 \begin_layout Plain Layout
2263
2264    function In_Arg (Data : Data_Type) return Arg_Type;
2265 \end_layout
2266
2267 \begin_layout Plain Layout
2268
2269 \end_layout
2270
2271 \begin_layout Plain Layout
2272
2273    function Out_Arg (Data : not null access Data_Type) return Arg_Type;
2274 \end_layout
2275
2276 \begin_layout Plain Layout
2277
2278 \end_layout
2279
2280 \begin_layout Plain Layout
2281
2282    function In_Out_Arg (Data : not null access Data_Type) return Arg_Type;
2283 \end_layout
2284
2285 \begin_layout Plain Layout
2286
2287 end Arg_Creators;
2288 \end_layout
2289
2290 \end_inset
2291
2292
2293 \end_layout
2294
2295 \begin_layout Standard
2296 \begin_inset ERT
2297 status open
2298
2299 \begin_layout Plain Layout
2300
2301
2302 \backslash
2303 end{frame}
2304 \end_layout
2305
2306 \end_inset
2307
2308
2309 \end_layout
2310
2311 \begin_layout Standard
2312 \begin_inset ERT
2313 status open
2314
2315 \begin_layout Plain Layout
2316
2317
2318 \backslash
2319 begin{frame}[fragile]
2320 \backslash
2321 frametitle{Kernel arguments III}
2322 \end_layout
2323
2324 \end_inset
2325
2326
2327 \end_layout
2328
2329 \begin_layout Standard
2330 \begin_inset listings
2331 inline false
2332 status open
2333
2334 \begin_layout Plain Layout
2335
2336    ...
2337 \end_layout
2338
2339 \begin_layout Plain Layout
2340
2341    package Matrix_Args is new CUDA.Compiler.Arg_Creators
2342 \end_layout
2343
2344 \begin_layout Plain Layout
2345
2346      (Data_Type => Ada.Numerics.Real_Arrays.Real_Matrix);
2347 \end_layout
2348
2349 \begin_layout Plain Layout
2350
2351    use Matrix_Args;
2352 \end_layout
2353
2354 \begin_layout Plain Layout
2355
2356 \end_layout
2357
2358 \begin_layout Plain Layout
2359
2360    Matrix : Ada.Numerics.Real_Arrays.Real_Matrix
2361 \end_layout
2362
2363 \begin_layout Plain Layout
2364
2365      := (1 ..
2366  N => (1 ..
2367  N => 0.0));
2368 \end_layout
2369
2370 \begin_layout Plain Layout
2371
2372    Arg    : CUDA.Compiler.Arg_Type
2373 \end_layout
2374
2375 \begin_layout Plain Layout
2376
2377      := In_Arg (Data => Matrix);
2378 \end_layout
2379
2380 \begin_layout Plain Layout
2381
2382 begin
2383 \end_layout
2384
2385 \begin_layout Plain Layout
2386
2387    ...
2388 \end_layout
2389
2390 \end_inset
2391
2392
2393 \end_layout
2394
2395 \begin_layout Standard
2396 \begin_inset ERT
2397 status open
2398
2399 \begin_layout Plain Layout
2400
2401
2402 \backslash
2403 end{frame}
2404 \end_layout
2405
2406 \end_inset
2407
2408
2409 \end_layout
2410
2411 \begin_layout Standard
2412 \begin_inset ERT
2413 status open
2414
2415 \begin_layout Plain Layout
2416
2417
2418 \backslash
2419 begin{frame}[fragile]
2420 \backslash
2421 frametitle{Kernel arguments IV}
2422 \end_layout
2423
2424 \end_inset
2425
2426
2427 \end_layout
2428
2429 \begin_layout Standard
2430
2431 \family typewriter
2432 \series bold
2433 Kernel signature
2434 \end_layout
2435
2436 \begin_layout Standard
2437 \begin_inset listings
2438 lstparams "language=C"
2439 inline false
2440 status open
2441
2442 \begin_layout Plain Layout
2443
2444 void mul(float* A, float* B, float* C)
2445 \end_layout
2446
2447 \end_inset
2448
2449
2450 \end_layout
2451
2452 \begin_layout Standard
2453
2454 \family typewriter
2455 \series bold
2456 CUDA/Ada call
2457 \end_layout
2458
2459 \begin_layout Standard
2460 \begin_inset listings
2461 inline false
2462 status open
2463
2464 \begin_layout Plain Layout
2465
2466    Func.Call 
2467 \end_layout
2468
2469 \begin_layout Plain Layout
2470
2471      (Args =>
2472 \end_layout
2473
2474 \begin_layout Plain Layout
2475
2476         (1 => In_Arg  (Data => A),
2477 \end_layout
2478
2479 \begin_layout Plain Layout
2480
2481          2 => In_Arg  (Data => B),
2482 \end_layout
2483
2484 \begin_layout Plain Layout
2485
2486          3 => Out_Arg (Data => C'Access)));
2487 \end_layout
2488
2489 \end_inset
2490
2491
2492 \end_layout
2493
2494 \begin_layout Standard
2495 \begin_inset ERT
2496 status open
2497
2498 \begin_layout Plain Layout
2499
2500
2501 \backslash
2502 end{frame}
2503 \end_layout
2504
2505 \end_inset
2506
2507
2508 \end_layout
2509
2510 \begin_layout Standard
2511 \begin_inset ERT
2512 status open
2513
2514 \begin_layout Plain Layout
2515
2516
2517 \backslash
2518 begin{frame}[fragile]
2519 \backslash
2520 frametitle{Error handling}
2521 \end_layout
2522
2523 \end_inset
2524
2525
2526 \end_layout
2527
2528 \begin_layout Block
2529 \begin_inset ERT
2530 status collapsed
2531
2532 \begin_layout Plain Layout
2533
2534 {
2535 \end_layout
2536
2537 \end_inset
2538
2539
2540 \family typewriter
2541 Functionality
2542 \family default
2543
2544 \begin_inset ERT
2545 status collapsed
2546
2547 \begin_layout Plain Layout
2548
2549 }
2550 \end_layout
2551
2552 \end_inset
2553
2554
2555 \end_layout
2556
2557 \begin_deeper
2558 \begin_layout Itemize
2559 Translation of CUDA errors to Ada exceptions
2560 \end_layout
2561
2562 \begin_layout Itemize
2563 Automated error checking of all low-level Thin-Binding calls
2564 \end_layout
2565
2566 \begin_layout Itemize
2567 Resolution of error code to error message 
2568 \end_layout
2569
2570 \end_deeper
2571 \begin_layout ExampleBlock
2572 \begin_inset ERT
2573 status collapsed
2574
2575 \begin_layout Plain Layout
2576
2577 {
2578 \end_layout
2579
2580 \end_inset
2581
2582 Requesting a nonexistent kernel 'Matrix_Mul'
2583 \begin_inset ERT
2584 status collapsed
2585
2586 \begin_layout Plain Layout
2587
2588 }
2589 \end_layout
2590
2591 \end_inset
2592
2593
2594 \end_layout
2595
2596 \begin_deeper
2597 \begin_layout Standard
2598 \begin_inset ERT
2599 status open
2600
2601 \begin_layout Plain Layout
2602
2603
2604 \backslash
2605 begin{verbatim}
2606 \end_layout
2607
2608 \begin_layout Plain Layout
2609
2610 Execution terminated by unhandled exception
2611 \end_layout
2612
2613 \begin_layout Plain Layout
2614
2615 Exception name: CUDA.CUDA_ERROR
2616 \end_layout
2617
2618 \begin_layout Plain Layout
2619
2620 Message: Could not get function Matrix_Mul (Not found)
2621 \end_layout
2622
2623 \begin_layout Plain Layout
2624
2625 Call stack traceback locations:
2626 \end_layout
2627
2628 \begin_layout Plain Layout
2629
2630 0x4079b6 0x40bc3a 0x406a4b 0x406299
2631 \end_layout
2632
2633 \begin_layout Plain Layout
2634
2635 0x7f159e4afc4b 0x405bd7
2636 \end_layout
2637
2638 \begin_layout Plain Layout
2639
2640
2641 \backslash
2642 end{verbatim}
2643 \end_layout
2644
2645 \end_inset
2646
2647
2648 \end_layout
2649
2650 \end_deeper
2651 \begin_layout Standard
2652 \begin_inset ERT
2653 status open
2654
2655 \begin_layout Plain Layout
2656
2657
2658 \backslash
2659 end{frame}
2660 \end_layout
2661
2662 \end_inset
2663
2664
2665 \end_layout
2666
2667 \begin_layout Standard
2668 \begin_inset ERT
2669 status open
2670
2671 \begin_layout Plain Layout
2672
2673
2674 \backslash
2675 begin{frame}[fragile]
2676 \backslash
2677 frametitle{Querying CUDA devices}
2678 \end_layout
2679
2680 \end_inset
2681
2682
2683 \end_layout
2684
2685 \begin_layout Block
2686 \begin_inset ERT
2687 status collapsed
2688
2689 \begin_layout Plain Layout
2690
2691 {
2692 \end_layout
2693
2694 \end_inset
2695
2696
2697 \family typewriter
2698 Functionality
2699 \family default
2700
2701 \begin_inset ERT
2702 status collapsed
2703
2704 \begin_layout Plain Layout
2705
2706 }
2707 \end_layout
2708
2709 \end_inset
2710
2711
2712 \end_layout
2713
2714 \begin_deeper
2715 \begin_layout Itemize
2716 Enumerate all available CUDA devices
2717 \end_layout
2718
2719 \begin_layout Itemize
2720 Query device properties like name, compute capability, etc.
2721 \end_layout
2722
2723 \end_deeper
2724 \begin_layout Separator
2725
2726 \end_layout
2727
2728 \begin_layout Standard
2729 \begin_inset listings
2730 inline false
2731 status open
2732
2733 \begin_layout Plain Layout
2734
2735 with Ada.Text_IO;
2736 \end_layout
2737
2738 \begin_layout Plain Layout
2739
2740 with CUDA.Driver; use CUDA.Driver;
2741 \end_layout
2742
2743 \begin_layout Plain Layout
2744
2745 \end_layout
2746
2747 \begin_layout Plain Layout
2748
2749 procedure Enum_Devices is
2750 \end_layout
2751
2752 \begin_layout Plain Layout
2753
2754    procedure Print_Name (Dev : Device_Type) is
2755 \end_layout
2756
2757 \begin_layout Plain Layout
2758
2759    begin
2760 \end_layout
2761
2762 \begin_layout Plain Layout
2763
2764       Ada.Text_IO.Put_Line (Name (Dev));
2765 \end_layout
2766
2767 \begin_layout Plain Layout
2768
2769    end Print_Name;
2770 \end_layout
2771
2772 \begin_layout Plain Layout
2773
2774 begin
2775 \end_layout
2776
2777 \begin_layout Plain Layout
2778
2779    Iterate (Process => Print_Name'Access);
2780 \end_layout
2781
2782 \begin_layout Plain Layout
2783
2784 end Enum_Devices;
2785 \end_layout
2786
2787 \end_inset
2788
2789
2790 \end_layout
2791
2792 \begin_layout Standard
2793 \begin_inset ERT
2794 status open
2795
2796 \begin_layout Plain Layout
2797
2798
2799 \backslash
2800 end{frame}
2801 \end_layout
2802
2803 \end_inset
2804
2805
2806 \end_layout
2807
2808 \begin_layout Section
2809 Conclusion
2810 \end_layout
2811
2812 \begin_layout Subsection
2813 Performance
2814 \end_layout
2815
2816 \begin_layout BeginFrame
2817 Performance
2818 \end_layout
2819
2820 \begin_layout Block
2821 \begin_inset ERT
2822 status collapsed
2823
2824 \begin_layout Plain Layout
2825
2826 {
2827 \end_layout
2828
2829 \end_inset
2830
2831
2832 \family typewriter
2833 Ambition
2834 \family default
2835
2836 \begin_inset ERT
2837 status collapsed
2838
2839 \begin_layout Plain Layout
2840
2841 }
2842 \end_layout
2843
2844 \end_inset
2845
2846
2847 \end_layout
2848
2849 \begin_deeper
2850 \begin_layout Quote
2851 Programs using CUDA/Ada should reach 
2852 \begin_inset Quotes eld
2853 \end_inset
2854
2855 native
2856 \begin_inset Quotes erd
2857 \end_inset
2858
2859  CUDA speed.
2860  The overhead imposed by Ada should be kept minimal.
2861 \end_layout
2862
2863 \end_deeper
2864 \begin_layout Separator
2865
2866 \end_layout
2867
2868 \begin_layout Block
2869 \begin_inset ERT
2870 status collapsed
2871
2872 \begin_layout Plain Layout
2873
2874 {
2875 \end_layout
2876
2877 \end_inset
2878
2879
2880 \family typewriter
2881 Measurement
2882 \family default
2883  methodology
2884 \begin_inset ERT
2885 status collapsed
2886
2887 \begin_layout Plain Layout
2888
2889 }
2890 \end_layout
2891
2892 \end_inset
2893
2894
2895 \end_layout
2896
2897 \begin_deeper
2898 \begin_layout Itemize
2899 \begin_inset ERT
2900 status collapsed
2901
2902 \begin_layout Plain Layout
2903
2904 <1->
2905 \end_layout
2906
2907 \end_inset
2908
2909 Cumulated runtime of 20 matrix multiplications (512 x 512)
2910 \end_layout
2911
2912 \begin_layout Itemize
2913 \begin_inset ERT
2914 status collapsed
2915
2916 \begin_layout Plain Layout
2917
2918 <2->
2919 \end_layout
2920
2921 \end_inset
2922
2923 Comparison of different implementations:
2924 \end_layout
2925
2926 \begin_deeper
2927 \begin_layout Itemize
2928 Ada (CPU)
2929 \end_layout
2930
2931 \begin_layout Itemize
2932 CUDA Runtime API
2933 \end_layout
2934
2935 \begin_layout Itemize
2936 CUDA Driver API
2937 \end_layout
2938
2939 \begin_layout Itemize
2940 CUDA/Ada
2941 \end_layout
2942
2943 \end_deeper
2944 \begin_layout Itemize
2945 \begin_inset ERT
2946 status collapsed
2947
2948 \begin_layout Plain Layout
2949
2950 <3->
2951 \end_layout
2952
2953 \end_inset
2954
2955 Same grid and block size for all CUDA implementations
2956 \end_layout
2957
2958 \end_deeper
2959 \begin_layout BeginFrame
2960 Performance II
2961 \end_layout
2962
2963 \begin_layout Standard
2964 \begin_inset Float table
2965 wide false
2966 sideways false
2967 status open
2968
2969 \begin_layout Plain Layout
2970 \align center
2971 \begin_inset Tabular
2972 <lyxtabular version="3" rows="7" columns="2">
2973 <features tabularvalignment="middle">
2974 <column alignment="center" valignment="top" width="0">
2975 <column alignment="center" valignment="top" width="0pt">
2976 <row>
2977 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
2978 \begin_inset Text
2979
2980 \begin_layout Plain Layout
2981 Processor
2982 \end_layout
2983
2984 \end_inset
2985 </cell>
2986 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
2987 \begin_inset Text
2988
2989 \begin_layout Plain Layout
2990 AMD Phenom II X4 940
2991 \end_layout
2992
2993 \end_inset
2994 </cell>
2995 </row>
2996 <row>
2997 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
2998 \begin_inset Text
2999
3000 \begin_layout Plain Layout
3001 Graphics Card
3002 \end_layout
3003
3004 \end_inset
3005 </cell>
3006 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3007 \begin_inset Text
3008
3009 \begin_layout Plain Layout
3010 GeForce GTX 560 Ti
3011 \end_layout
3012
3013 \end_inset
3014 </cell>
3015 </row>
3016 <row>
3017 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3018 \begin_inset Text
3019
3020 \begin_layout Plain Layout
3021 Operating System
3022 \end_layout
3023
3024 \end_inset
3025 </cell>
3026 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3027 \begin_inset Text
3028
3029 \begin_layout Plain Layout
3030 Debian Linux 6.0
3031 \end_layout
3032
3033 \end_inset
3034 </cell>
3035 </row>
3036 <row>
3037 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3038 \begin_inset Text
3039
3040 \begin_layout Plain Layout
3041 Kernel
3042 \end_layout
3043
3044 \end_inset
3045 </cell>
3046 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3047 \begin_inset Text
3048
3049 \begin_layout Plain Layout
3050  2.6.32-5-amd64
3051 \end_layout
3052
3053 \end_inset
3054 </cell>
3055 </row>
3056 <row>
3057 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3058 \begin_inset Text
3059
3060 \begin_layout Plain Layout
3061 Ada Compiler
3062 \end_layout
3063
3064 \end_inset
3065 </cell>
3066 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3067 \begin_inset Text
3068
3069 \begin_layout Plain Layout
3070 FSF GNAT 4.4.5
3071 \end_layout
3072
3073 \end_inset
3074 </cell>
3075 </row>
3076 <row>
3077 <cell alignment="center" valignment="top" topline="true" leftline="true" usebox="none">
3078 \begin_inset Text
3079
3080 \begin_layout Plain Layout
3081 NVIDIA Graphics Driver
3082 \end_layout
3083
3084 \end_inset
3085 </cell>
3086 <cell alignment="center" valignment="top" topline="true" leftline="true" rightline="true" usebox="none">
3087 \begin_inset Text
3088
3089 \begin_layout Plain Layout
3090 270.41.19, Linux 64-bit
3091 \end_layout
3092
3093 \end_inset
3094 </cell>
3095 </row>
3096 <row>
3097 <cell alignment="center" valignment="top" topline="true" bottomline="true" leftline="true" usebox="none">
3098 \begin_inset Text
3099
3100 \begin_layout Plain Layout
3101 NVIDIA CUDA Toolkit
3102 \end_layout
3103
3104 \end_inset
3105 </cell>
3106 <cell alignment="center" valignment="top" topline="true" bottomline="true" leftline="true" rightline="true" usebox="none">
3107 \begin_inset Text
3108
3109 \begin_layout Plain Layout
3110 4.0.17, Linux 64-bit
3111 \end_layout
3112
3113 \end_inset
3114 </cell>
3115 </row>
3116 </lyxtabular>
3117
3118 \end_inset
3119
3120
3121 \begin_inset Caption
3122
3123 \begin_layout Plain Layout
3124 Test system specs
3125 \end_layout
3126
3127 \end_inset
3128
3129
3130 \end_layout
3131
3132 \end_inset
3133
3134
3135 \end_layout
3136
3137 \begin_layout EndFrame
3138
3139 \end_layout
3140
3141 \begin_layout BeginFrame
3142 Benchmarking results
3143 \end_layout
3144
3145 \begin_layout Standard
3146 \align center
3147 \begin_inset Graphics
3148         filename performance-chart.pdf
3149         scale 76
3150
3151 \end_inset
3152
3153
3154 \end_layout
3155
3156 \begin_layout EndFrame
3157
3158 \end_layout
3159
3160 \begin_layout BeginFrame
3161 Performance III
3162 \end_layout
3163
3164 \begin_layout Block
3165 \begin_inset ERT
3166 status collapsed
3167
3168 \begin_layout Plain Layout
3169
3170 {
3171 \end_layout
3172
3173 \end_inset
3174
3175
3176 \family typewriter
3177 Interpretation
3178 \family default
3179
3180 \begin_inset ERT
3181 status collapsed
3182
3183 \begin_layout Plain Layout
3184
3185 }
3186 \end_layout
3187
3188 \end_inset
3189
3190
3191 \end_layout
3192
3193 \begin_deeper
3194 \begin_layout Itemize
3195 \begin_inset ERT
3196 status collapsed
3197
3198 \begin_layout Plain Layout
3199
3200 <1->
3201 \end_layout
3202
3203 \end_inset
3204
3205 Native Ada (CPU) is slow: CPU vs.
3206  GPU/CUDA
3207 \end_layout
3208
3209 \begin_layout Itemize
3210 \begin_inset ERT
3211 status collapsed
3212
3213 \begin_layout Plain Layout
3214
3215 <2->
3216 \end_layout
3217
3218 \end_inset
3219
3220 CUDA/Ada is faster than CUDA Runtime API:
3221 \end_layout
3222
3223 \begin_deeper
3224 \begin_layout Itemize
3225 CUDA Runtime API generates management as well as kernel launch code, etc.
3226 \end_layout
3227
3228 \begin_layout Itemize
3229 CUDA/Ada performs bare minimum of management operations
3230 \end_layout
3231
3232 \end_deeper
3233 \begin_layout Itemize
3234 \begin_inset ERT
3235 status collapsed
3236
3237 \begin_layout Plain Layout
3238
3239 <3->
3240 \end_layout
3241
3242 \end_inset
3243
3244 CUDA/Ada is negligibly slower than CUDA Driver API
3245 \end_layout
3246
3247 \begin_layout Itemize
3248 \begin_inset ERT
3249 status collapsed
3250
3251 \begin_layout Plain Layout
3252
3253 <4->
3254 \end_layout
3255
3256 \end_inset
3257
3258 No visible performance penalty
3259 \end_layout
3260
3261 \end_deeper
3262 \begin_layout EndFrame
3263
3264 \end_layout
3265
3266 \begin_layout Subsection
3267 Review
3268 \end_layout
3269
3270 \begin_layout BeginFrame
3271 Review
3272 \end_layout
3273
3274 \begin_layout Block
3275 \begin_inset ERT
3276 status collapsed
3277
3278 \begin_layout Plain Layout
3279
3280 {
3281 \end_layout
3282
3283 \end_inset
3284
3285 Assessment of results
3286 \begin_inset ERT
3287 status collapsed
3288
3289 \begin_layout Plain Layout
3290
3291 }
3292 \end_layout
3293
3294 \end_inset
3295
3296
3297 \end_layout
3298
3299 \begin_deeper
3300 \begin_layout Itemize
3301 \begin_inset ERT
3302 status collapsed
3303
3304 \begin_layout Plain Layout
3305
3306 <1->
3307 \end_layout
3308
3309 \end_inset
3310
3311 Thick-Binding provides easy usage of CUDA from Ada
3312 \end_layout
3313
3314 \begin_layout Itemize
3315 \begin_inset ERT
3316 status collapsed
3317
3318 \begin_layout Plain Layout
3319
3320 <2->
3321 \end_layout
3322
3323 \end_inset
3324
3325 High level abstractions go well with other Ada language constructs
3326 \end_layout
3327
3328 \begin_layout Itemize
3329 \begin_inset ERT
3330 status collapsed
3331
3332 \begin_layout Plain Layout
3333
3334 <3->
3335 \end_layout
3336
3337 \end_inset
3338
3339 Simple Autoinitialization of CUDA via 
3340 \family typewriter
3341 Autoinit
3342 \family default
3343  package
3344 \end_layout
3345
3346 \begin_layout Itemize
3347 \begin_inset ERT
3348 status collapsed
3349
3350 \begin_layout Plain Layout
3351
3352 <4->
3353 \end_layout
3354
3355 \end_inset
3356
3357 JIT compilation of kernels provides flexibilty
3358 \end_layout
3359
3360 \begin_layout Itemize
3361 \begin_inset ERT
3362 status collapsed
3363
3364 \begin_layout Plain Layout
3365
3366 <5->
3367 \end_layout
3368
3369 \end_inset
3370
3371 Speed of CUDA/Ada is excellent
3372 \end_layout
3373
3374 \begin_layout Itemize
3375 \begin_inset ERT
3376 status collapsed
3377
3378 \begin_layout Plain Layout
3379
3380 <6->
3381 \end_layout
3382
3383 \end_inset
3384
3385 Automated Thin-Binding generation allows for easy integration of future
3386  CUDA API changes
3387 \end_layout
3388
3389 \begin_layout Itemize
3390 \begin_inset ERT
3391 status collapsed
3392
3393 \begin_layout Plain Layout
3394
3395 <7->
3396 \end_layout
3397
3398 \end_inset
3399
3400 Paper documents how to write a modern, maintainable language binding for
3401  Ada in general
3402 \end_layout
3403
3404 \end_deeper
3405 \begin_layout EndFrame
3406
3407 \end_layout
3408
3409 \begin_layout BeginFrame
3410 Review II
3411 \end_layout
3412
3413 \begin_layout Block
3414 \begin_inset ERT
3415 status collapsed
3416
3417 \begin_layout Plain Layout
3418
3419 {
3420 \end_layout
3421
3422 \end_inset
3423
3424 Deliverables
3425 \begin_inset ERT
3426 status collapsed
3427
3428 \begin_layout Plain Layout
3429
3430 }
3431 \end_layout
3432
3433 \end_inset
3434
3435
3436 \end_layout
3437
3438 \begin_deeper
3439 \begin_layout Itemize
3440 Paper
3441 \end_layout
3442
3443 \begin_layout Itemize
3444 Slides
3445 \end_layout
3446
3447 \begin_layout Itemize
3448 Source code (GPLv3+)
3449 \family typewriter
3450
3451 \begin_inset Newline newline
3452 \end_inset
3453
3454
3455 \family default
3456
3457 \begin_inset Flex URL
3458 status collapsed
3459
3460 \begin_layout Plain Layout
3461
3462 http://git.codelabs.ch/?p=cuda-ada.git
3463 \end_layout
3464
3465 \end_inset
3466
3467
3468 \family typewriter
3469
3470 \begin_inset Newline newline
3471 \end_inset
3472
3473 git clone http://git.codelabs.ch/git/cuda-ada.git
3474 \end_layout
3475
3476 \begin_layout Itemize
3477 Website with all documents and additional information: 
3478 \begin_inset Flex URL
3479 status collapsed
3480
3481 \begin_layout Plain Layout
3482
3483 http://www.codelabs.ch/cuda-ada/
3484 \end_layout
3485
3486 \end_inset
3487
3488
3489 \end_layout
3490
3491 \end_deeper
3492 \begin_layout EndFrame
3493
3494 \end_layout
3495
3496 \begin_layout Subsection
3497 Outlook
3498 \end_layout
3499
3500 \begin_layout BeginFrame
3501 Outlook
3502 \end_layout
3503
3504 \begin_layout Block
3505 \begin_inset ERT
3506 status collapsed
3507
3508 \begin_layout Plain Layout
3509
3510 {
3511 \end_layout
3512
3513 \end_inset
3514
3515 CUDA/Ada
3516 \begin_inset ERT
3517 status collapsed
3518
3519 \begin_layout Plain Layout
3520
3521 }
3522 \end_layout
3523
3524 \end_inset
3525
3526
3527 \end_layout
3528
3529 \begin_deeper
3530 \begin_layout Itemize
3531 Implement additional features:
3532 \end_layout
3533
3534 \begin_deeper
3535 \begin_layout Itemize
3536 Abstractions for kernel generation (e.g.
3537  element-wise)
3538 \end_layout
3539
3540 \begin_layout Itemize
3541 Kernel signature verification on 
3542 \family typewriter
3543 Call
3544 \end_layout
3545
3546 \begin_layout Itemize
3547 Multi-Device support
3548 \end_layout
3549
3550 \begin_layout Itemize
3551 ...
3552 \end_layout
3553
3554 \end_deeper
3555 \begin_layout Itemize
3556 Announce project to Ada community
3557 \end_layout
3558
3559 \end_deeper
3560 \begin_layout Separator
3561
3562 \end_layout
3563
3564 \begin_layout Block
3565 \begin_inset ERT
3566 status collapsed
3567
3568 \begin_layout Plain Layout
3569
3570 {
3571 \end_layout
3572
3573 \end_inset
3574
3575 Ada and CUDA
3576 \begin_inset ERT
3577 status collapsed
3578
3579 \begin_layout Plain Layout
3580
3581 }
3582 \end_layout
3583
3584 \end_inset
3585
3586
3587 \end_layout
3588
3589 \begin_deeper
3590 \begin_layout Itemize
3591 Study NVIDIA's LLVM-based CUDA compiler (announced, not yet released)
3592 \end_layout
3593
3594 \end_deeper
3595 \begin_layout EndFrame
3596
3597 \end_layout
3598
3599 \begin_layout BeginFrame
3600 Questions
3601 \end_layout
3602
3603 \begin_layout Standard
3604 \align center
3605
3606 \size giant
3607 Thank you for your attention!
3608 \end_layout
3609
3610 \begin_layout EndFrame
3611
3612 \end_layout
3613
3614 \end_body
3615 \end_document