1nvcc_PROBE() {
2    if [ -z "$REAL_NVCC" ]; then
3        echo "nvcc is not available"
4    elif [ -z "$REAL_CUOBJDUMP" ]; then
5        echo "cuobjdump is not available"
6    fi
7}
8
9nvcc_SETUP() {
10    # Test code using only c++ (option --x c++). Faster than compiling cuda.
11    cat <<EOF > test_cpp.cu
12#ifndef NUM
13#define NUM 10000
14#endif
15
16void caller() {
17  for (int i = 0; i < NUM; ++i);
18}
19EOF
20
21    # Option files to modify the define.
22    cat <<EOF >test1.optf
23-DNUM=1
24EOF
25    cat <<EOF >test2.optf
26-DNUM=2
27EOF
28
29    # Test code using cuda.
30    cat <<EOF >test_cuda.cu
31#ifndef NUM
32#define NUM 10000
33#endif
34
35__global__
36void add(int *a, int *b) {
37  int i = blockIdx.x;
38  if (i < NUM) {
39    b[i] = 2 * a[i];
40  }
41}
42
43void caller() {
44  add<<<NUM, 1>>>(NULL,NULL);
45}
46EOF
47}
48
49nvcc_tests() {
50    # Reference file testing was not successful due to different "fatbin" data.
51    # Another source of differences are the temporary files created by nvcc;
52    # that can be avoided by using the options "--keep --keep-dir ./keep". So
53    # instead of comparing the binary object files, we compare the dumps of
54    # "cuobjdump -all -elf -symbols -ptx -sass test1.o".
55    nvcc_opts_cpp="-Wno-deprecated-gpu-targets -c --x c++"
56    nvcc_opts_cuda="-Wno-deprecated-gpu-targets -c"
57    nvcc_opts_gpu1="--generate-code arch=compute_50,code=compute_50"
58    nvcc_opts_gpu2="--generate-code arch=compute_52,code=sm_52"
59    ccache_nvcc_cpp="$CCACHE $REAL_NVCC $nvcc_opts_cpp"
60    ccache_nvcc_cuda="$CCACHE $REAL_NVCC $nvcc_opts_cuda"
61    cuobjdump="$REAL_CUOBJDUMP -all -elf -symbols -ptx -sass"
62
63    # -------------------------------------------------------------------------
64    TEST "Simple mode"
65
66    $REAL_NVCC $nvcc_opts_cpp -o reference_test1.o test_cpp.cu
67
68    # First compile.
69    $ccache_nvcc_cpp test_cpp.cu
70    expect_stat 'cache hit (preprocessed)' 0
71    expect_stat 'cache miss' 1
72    expect_stat 'files in cache' 1
73    expect_equal_files reference_test1.o test_cpp.o
74
75    $ccache_nvcc_cpp test_cpp.cu
76    expect_stat 'cache hit (preprocessed)' 1
77    expect_stat 'cache miss' 1
78    expect_stat 'files in cache' 1
79    expect_equal_files reference_test1.o test_cpp.o
80
81    # -------------------------------------------------------------------------
82    TEST "Different GPU architectures"
83
84    $REAL_NVCC $nvcc_opts_cuda                 -o reference_test1.o test_cuda.cu
85    $REAL_NVCC $nvcc_opts_cuda $nvcc_opts_gpu1 -o reference_test2.o test_cuda.cu
86    $REAL_NVCC $nvcc_opts_cuda $nvcc_opts_gpu2 -o reference_test3.o test_cuda.cu
87    $cuobjdump reference_test1.o > reference_test1.dump
88    $cuobjdump reference_test2.o > reference_test2.dump
89    $cuobjdump reference_test3.o > reference_test3.dump
90    expect_different_files reference_test1.dump reference_test2.dump
91    expect_different_files reference_test1.dump reference_test3.dump
92    expect_different_files reference_test2.dump reference_test3.dump
93
94    $ccache_nvcc_cuda test_cuda.cu
95    expect_stat 'cache hit (preprocessed)' 0
96    expect_stat 'cache miss' 1
97    expect_stat 'files in cache' 1
98    $cuobjdump test_cuda.o > test1.dump
99    expect_equal_files reference_test1.dump test1.dump
100
101    # Other GPU.
102    $ccache_nvcc_cuda $nvcc_opts_gpu1 test_cuda.cu
103    expect_stat 'cache hit (preprocessed)' 0
104    expect_stat 'cache miss' 2
105    expect_stat 'files in cache' 2
106    $cuobjdump test_cuda.o > test1.dump
107    expect_equal_files reference_test2.dump test1.dump
108
109    $ccache_nvcc_cuda $nvcc_opts_gpu1 test_cuda.cu
110    expect_stat 'cache hit (preprocessed)' 1
111    expect_stat 'cache miss' 2
112    expect_stat 'files in cache' 2
113    $cuobjdump test_cuda.o > test1.dump
114    expect_equal_files reference_test2.dump test1.dump
115
116    # Another GPU.
117    $ccache_nvcc_cuda $nvcc_opts_gpu2 test_cuda.cu
118    expect_stat 'cache hit (preprocessed)' 1
119    expect_stat 'cache miss' 3
120    expect_stat 'files in cache' 3
121    $cuobjdump test_cuda.o > test1.dump
122    expect_equal_files reference_test3.dump test1.dump
123
124    $ccache_nvcc_cuda $nvcc_opts_gpu2 test_cuda.cu
125    expect_stat 'cache hit (preprocessed)' 2
126    expect_stat 'cache miss' 3
127    expect_stat 'files in cache' 3
128    $cuobjdump test_cuda.o > test1.dump
129    expect_equal_files reference_test3.dump test1.dump
130
131    # -------------------------------------------------------------------------
132    TEST "Different defines"
133
134    $REAL_NVCC $nvcc_opts_cpp            -o reference_test1.o test_cpp.cu
135    $REAL_NVCC $nvcc_opts_cpp -DNUM=10   -o reference_test2.o test_cpp.cu
136    expect_different_files reference_test1.o reference_test2.o
137
138    $ccache_nvcc_cpp test_cpp.cu
139    expect_stat 'cache hit (preprocessed)' 0
140    expect_stat 'cache miss' 1
141    expect_stat 'files in cache' 1
142    expect_equal_files reference_test1.o test_cpp.o
143
144    # Specified define, but unused. Can only be found by preprocessed mode.
145    $ccache_nvcc_cpp -DDUMMYENV=1 test_cpp.cu
146    expect_stat "cache hit (preprocessed)" 1
147    expect_stat 'cache miss' 1
148    expect_stat 'files in cache' 1
149    expect_equal_files reference_test1.o test_cpp.o
150
151    # Specified used define.
152    $ccache_nvcc_cpp -DNUM=10 test_cpp.cu
153    expect_stat "cache hit (preprocessed)" 1
154    expect_stat 'cache miss' 2
155    expect_stat 'files in cache' 2
156    expect_equal_files reference_test2.o test_cpp.o
157
158    $ccache_nvcc_cpp -DNUM=10 test_cpp.cu
159    expect_stat 'cache hit (preprocessed)' 2
160    expect_stat 'cache miss' 2
161    expect_stat 'files in cache' 2
162    expect_equal_files reference_test2.o test_cpp.o
163
164    # -------------------------------------------------------------------------
165    TEST "Option file"
166
167    $REAL_NVCC $nvcc_opts_cpp -optf test1.optf -o reference_test1.o test_cpp.cu
168    $REAL_NVCC $nvcc_opts_cpp -optf test2.optf -o reference_test2.o test_cpp.cu
169    expect_different_files reference_test1.o reference_test2.o
170
171    $ccache_nvcc_cpp -optf test1.optf test_cpp.cu
172    expect_stat 'cache hit (preprocessed)' 0
173    expect_stat 'cache miss' 1
174    expect_stat 'files in cache' 1
175    expect_equal_files reference_test1.o test_cpp.o
176
177    $ccache_nvcc_cpp -optf test1.optf test_cpp.cu
178    expect_stat 'cache hit (preprocessed)' 1
179    expect_stat 'cache miss' 1
180    expect_stat 'files in cache' 1
181    expect_equal_files reference_test1.o test_cpp.o
182
183    $ccache_nvcc_cpp -optf test2.optf test_cpp.cu
184    expect_stat 'cache hit (preprocessed)' 1
185    expect_stat 'cache miss' 2
186    expect_stat 'files in cache' 2
187    expect_equal_files reference_test2.o test_cpp.o
188
189    $ccache_nvcc_cpp -optf test2.optf test_cpp.cu
190    expect_stat 'cache hit (preprocessed)' 2
191    expect_stat 'cache miss' 2
192    expect_stat 'files in cache' 2
193    expect_equal_files reference_test2.o test_cpp.o
194
195    # -------------------------------------------------------------------------
196    TEST "Option --compiler-bindir"
197
198    $REAL_NVCC $nvcc_opts_cpp --compiler-bindir $REAL_COMPILER_BIN \
199      -o reference_test1.o test_cpp.cu
200
201    # First compile.
202    $ccache_nvcc_cpp --compiler-bindir $REAL_COMPILER_BIN test_cpp.cu
203    expect_stat 'cache hit (preprocessed)' 0
204    expect_stat 'cache miss' 1
205    expect_stat 'files in cache' 1
206    expect_equal_files reference_test1.o test_cpp.o
207
208    $ccache_nvcc_cpp --compiler-bindir $REAL_COMPILER_BIN test_cpp.cu
209    expect_stat 'cache hit (preprocessed)' 1
210    expect_stat 'cache miss' 1
211    expect_stat 'files in cache' 1
212    expect_equal_files reference_test1.o test_cpp.o
213
214    # -------------------------------------------------------------------------
215    TEST "Option -ccbin"
216
217    $REAL_NVCC $nvcc_opts_cpp -ccbin $REAL_COMPILER_BIN \
218      -o reference_test1.o test_cpp.cu
219
220    # First compile.
221    $ccache_nvcc_cpp -ccbin $REAL_COMPILER_BIN test_cpp.cu
222    expect_stat 'cache hit (preprocessed)' 0
223    expect_stat 'cache miss' 1
224    expect_stat 'files in cache' 1
225    expect_equal_files reference_test1.o test_cpp.o
226
227    $ccache_nvcc_cpp -ccbin $REAL_COMPILER_BIN test_cpp.cu
228    expect_stat 'cache hit (preprocessed)' 1
229    expect_stat 'cache miss' 1
230    expect_stat 'files in cache' 1
231    expect_equal_files reference_test1.o test_cpp.o
232
233    # -------------------------------------------------------------------------
234    TEST "Option --output-directory"
235
236    $REAL_NVCC $nvcc_opts_cpp --output-directory . \
237      -o reference_test1.o test_cpp.cu
238
239    # First compile.
240    $ccache_nvcc_cpp --output-directory . test_cpp.cu
241    expect_stat 'cache hit (preprocessed)' 0
242    expect_stat 'cache miss' 1
243    expect_stat 'files in cache' 1
244    expect_equal_files reference_test1.o test_cpp.o
245
246    $ccache_nvcc_cpp --output-directory . test_cpp.cu
247    expect_stat 'cache hit (preprocessed)' 1
248    expect_stat 'cache miss' 1
249    expect_stat 'files in cache' 1
250    expect_equal_files reference_test1.o test_cpp.o
251
252    # -------------------------------------------------------------------------
253    TEST "Option -odir"
254
255    $REAL_NVCC $nvcc_opts_cpp -odir . -o reference_test1.o test_cpp.cu
256
257    # First compile.
258    $ccache_nvcc_cpp -odir . test_cpp.cu
259    expect_stat 'cache hit (preprocessed)' 0
260    expect_stat 'cache miss' 1
261    expect_stat 'files in cache' 1
262    expect_equal_files reference_test1.o test_cpp.o
263
264    $ccache_nvcc_cpp -odir . test_cpp.cu
265    expect_stat 'cache hit (preprocessed)' 1
266    expect_stat 'cache miss' 1
267    expect_stat 'files in cache' 1
268    expect_equal_files reference_test1.o test_cpp.o
269}
270
271SUITE_nvcc_PROBE() {
272    nvcc_PROBE
273}
274
275SUITE_nvcc_SETUP() {
276    nvcc_SETUP
277}
278
279SUITE_nvcc() {
280    nvcc_tests
281}
282