Tests/Cuda/Complex/file3.cu


1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48


#include <iostream>

#include "file1.h"
#include "file2.h"

result_type __device__ file1_func(int x);
result_type_dynamic __device__ file2_func(int x);

static __global__ void file3_kernel(result_type* r, int x)
{
  *r = file1_func(x);
  result_type_dynamic rd = file2_func(x);
}

int file3_launch_kernel(int x)
{
  result_type* r;
  cudaError_t err = cudaMallocManaged(&r, sizeof(result_type));
  if (err != cudaSuccess) {
    std::cerr << "file3_launch_kernel: cudaMallocManaged failed: "
              << cudaGetErrorString(err) << std::endl;
    return x;
  }

  file3_kernel<<<1, 1>>>(r, x);
  err = cudaGetLastError();
  if (err != cudaSuccess) {
    std::cerr << "file3_kernel [SYNC] failed: " << cudaGetErrorString(err)
              << std::endl;
    return x;
  }
  err = cudaDeviceSynchronize();
  if (err != cudaSuccess) {
    std::cerr << "file3_kernel [ASYNC] failed: "
              << cudaGetErrorString(cudaGetLastError()) << std::endl;
    return x;
  }
  int result = r->sum;
  err = cudaFree(r);
  if (err != cudaSuccess) {
    std::cerr << "file3_launch_kernel: cudaFree failed: "
              << cudaGetErrorString(err) << std::endl;
    return x;
  }

  return result;
}