#include <iostream>

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

#ifdef _WIN32
#define EXPORT __declspec(dllexport)
#define IMPORT __declspec(dllimport)
#else
#define EXPORT
#define IMPORT
#endif

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

IMPORT void __host__ cuda_dynamic_lib_func();

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

EXPORT int mixed_launch_kernel(int x)
{
  cuda_dynamic_lib_func();

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

  mixed_kernel <<<1,1>>> (r,x);
  err = cudaGetLastError();
  if(err != cudaSuccess)
    {
    std::cerr << "mixed_kernel [SYNC] failed: "
              << cudaGetErrorString(err) << std::endl;
    return x;
    }
  err = cudaDeviceSynchronize();
  if(err != cudaSuccess)
    {
    std::cerr << "mixed_kernel [ASYNC] failed: "
              << cudaGetErrorString(cudaGetLastError()) << std::endl;
    return x;
    }

  int result = r->sum;
  err = cudaFree(r);
  if(err != cudaSuccess)
    {
    std::cerr << "mixed_launch_kernel: cudaFree failed: "
              << cudaGetErrorString(err) << std::endl;
    return x;
    }

  return result;
}