sample.cpp File Reference

#include <cassert>
#include <chrono>
#include <cstdint>
#include <cstring>
#include <fstream>
#include <iostream>
#include "Logger/Logger.h"
#include "cuda.h"

Include dependency graph for sample.cpp:

Go to the source code of this file.

Classes
struct	measure< TimeT >

Functions
void	checkCudaErrors (CUresult err)
int	main (int argc, char **argv)
	main - Program entry point More...

Function Documentation

void checkCudaErrors

(

CUresult

err

)

Definition at line 38 of file sample.cpp.

Referenced by copy_to_nvidia_gpu(), Executor::interrupt(), TableFunctionExecutionContext::launchGpuCode(), main(), QueryEngine::QueryEngine(), and QueryEngine::~QueryEngine().

                                   {
  if (err != CUDA_SUCCESS) {
    std::cout << err << std::endl;
  }
  assert(err == CUDA_SUCCESS);
}

Here is the caller graph for this function:

int main	(	int	argc,
		char **	argv
	)

main - Program entry point

Definition at line 46 of file sample.cpp.

References checkCudaErrors(), measure< TimeT >::execution(), logger::INFO, LOG, anonymous_namespace{Utm.h}::N, and run_benchmark_import::result.

                                {
  CUdevice device;
  CUmodule cudaModule;
  CUcontext context;
  CUfunction function;

  // CUDA initialization
  checkCudaErrors(cuInit(0));
  checkCudaErrors(cuDeviceGet(&device, 0));

  std::ifstream t("kernel.ptx");
  std::string str((std::istreambuf_iterator<char>(t)), std::istreambuf_iterator<char>());

  // Create driver context
  checkCudaErrors(cuCtxCreate(&context, 0, device));

  // Create module for object
  checkCudaErrors(cuModuleLoadDataEx(&cudaModule, str.c_str(), 0, 0, 0));

  // Get kernel function
  checkCudaErrors(cuModuleGetFunction(&function, cudaModule, "kernel"));

  int64_t N = 1000000000LL;
  int8_t* byte_stream_col_0 = new int8_t[N];
  memset(byte_stream_col_0, 42, N);

  CUdeviceptr devBufferA;
  checkCudaErrors(cuMemAlloc(&devBufferA, sizeof(int8_t) * N));
  checkCudaErrors(cuMemcpyHtoD(devBufferA, byte_stream_col_0, sizeof(int8_t) * N));

  CUdeviceptr devBufferAA;
  checkCudaErrors(cuMemAlloc(&devBufferAA, sizeof(CUdeviceptr)));
  checkCudaErrors(cuMemcpyHtoD(devBufferAA, &devBufferA, sizeof(CUdeviceptr)));

  unsigned blockSizeX = 128;
  unsigned blockSizeY = 1;
  unsigned blockSizeZ = 1;
  unsigned gridSizeX = 128;
  unsigned gridSizeY = 1;
  unsigned gridSizeZ = 1;

  CUdeviceptr devBufferB;
  int64_t* result_vec = new int64_t[blockSizeX * gridSizeX * sizeof(int64_t)];
  checkCudaErrors(cuMemAlloc(&devBufferB, blockSizeX * gridSizeX * sizeof(int64_t)));

  CUdeviceptr devBufferN;
  int64_t row_count = N;
  checkCudaErrors(cuMemAlloc(&devBufferN, sizeof(int64_t)));
  checkCudaErrors(cuMemcpyHtoD(devBufferN, &row_count, sizeof(int64_t)));

  CUdeviceptr devBufferI;
  int64_t init_agg_val = 0;
  checkCudaErrors(cuMemAlloc(&devBufferI, sizeof(int64_t)));
  checkCudaErrors(cuMemcpyHtoD(devBufferI, &init_agg_val, sizeof(int64_t)));

  void* KernelParams[] = {&devBufferAA, &devBufferN, &devBufferI, &devBufferB};

  LOG(INFO) << measure<std::chrono::microseconds>::execution([&]() {
    checkCudaErrors(cuLaunchKernel(function,
                                   gridSizeX,
                                   gridSizeY,
                                   gridSizeZ,
                                   blockSizeX,
                                   blockSizeY,
                                   blockSizeZ,
                                   0,
                                   NULL,
                                   KernelParams,
                                   NULL));
    checkCudaErrors(
        cuMemcpyDtoH(result_vec, devBufferB, blockSizeX * gridSizeX * sizeof(int64_t)));
  });

  int64_t result = 0;
  for (size_t i = 0; i < blockSizeX * gridSizeX; ++i) {
    result += result_vec[i];
  }
  std::cout << result << std::endl;

  delete[] result_vec;
  delete[] byte_stream_col_0;

  // Clean-up
  checkCudaErrors(cuMemFree(devBufferA));
  checkCudaErrors(cuMemFree(devBufferAA));
  checkCudaErrors(cuMemFree(devBufferB));
  checkCudaErrors(cuMemFree(devBufferN));
  checkCudaErrors(cuModuleUnload(cudaModule));
  checkCudaErrors(cuCtxDestroy(context));

  return 0;
}

Here is the call graph for this function: