Ziad Bkh Ziad Bkh - 5 months ago 46
Linux Question

CUDA : error: "transfer of control bypasses initialization of" when creating thrust::device_ptr

I have this code line in my Cuda - C application :

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/sort.h>
#include <stdio.h>

#include <time.h>

#include <device_functions.h>

int main()
{
const int size = 32;
unsigned int * dev_ips_range_end;
unsigned int * ips_range_end = new unsigned int[size];
for (int i = 0; i < size; i++)
ips_range_end[i] = i;



cudaError_t cudaStatus;
cudaStatus = cudaSetDevice(0);
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "cudaSetDevice failed! Do you have a CUDA-capable GPU installed?");
goto Error;
}
cudaStatus = cudaMalloc((void**)&dev_ips_range_end, size * sizeof(int));
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "Problem !");
goto Error;
}

// Copy input vectors from host memory to GPU buffers.
cudaStatus = cudaMemcpy(dev_ips_range_end, ips_range_end, size * sizeof(unsigned char), cudaMemcpyHostToDevice);
if (cudaStatus != cudaSuccess) {
fprintf(stderr, "Problem !");
goto Error;
}

thrust::device_ptr<unsigned int> dev_ips_range_end_ptr(dev_ips_range_end);
thrust::inclusive_scan(dev_ips_range_end_ptr, dev_ips_range_end_ptr + size, dev_ips_range_end_ptr);
return 0;

Error:
cudaFree(dev_ips_range_end);
}


here is the command I used and the output:

[Test]$ nvcc -I/usr/local/cuda/include -L/usr/local/cuda/lib kernel.cu -o test.run
kernel.cu(27): error: transfer of control bypasses initialization of:
variable "dev_ips_range_end_ptr"
(42): here

kernel.cu(32): error: transfer of control bypasses initialization of:
variable "dev_ips_range_end_ptr"
(42): here

kernel.cu(39): error: transfer of control bypasses initialization of:
variable "dev_ips_range_end_ptr"
(42): here

3 errors detected in the compilation of "/tmp/tmpxft_000022ad_00000000-9_kernel.cpp1.ii".

the same code is working without any problem in visual studio on windows.
how to solve this issue ?

Answer

Some people might tell you that the use of goto in C/C++ isn't a great idea. But to avoid arguments, and allow you to keep the same code structure, you can declare your thrust device pointer at the top of your program (before any goto statements) and then set the pointer value when you are ready to use it, like this:

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/sort.h>
#include <stdio.h>

#include <time.h>

#include <device_functions.h>

int main()
{
    const int size = 32;
    unsigned int * dev_ips_range_end;
    unsigned int * ips_range_end = new unsigned int[size];
    for (int i = 0; i < size; i++)
        ips_range_end[i] = i;



    thrust::device_ptr<unsigned int> dev_ips_range_end_ptr;
cudaError_t cudaStatus;
    cudaStatus = cudaSetDevice(0);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaSetDevice failed!  Do you have a CUDA-capable GPU installed?");
        goto Error;
    }
    cudaStatus = cudaMalloc((void**)&dev_ips_range_end, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "Problem !");
        goto Error;
    }

    // Copy input vectors from host memory to GPU buffers.
    cudaStatus = cudaMemcpy(dev_ips_range_end, ips_range_end, size * sizeof(unsigned char), cudaMemcpyHostToDevice);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "Problem !");
        goto Error;
    }

    dev_ips_range_end_ptr = thrust::device_pointer_cast(dev_ips_range_end);
    thrust::inclusive_scan(dev_ips_range_end_ptr, dev_ips_range_end_ptr + size, dev_ips_range_end_ptr);
    return 0;

Error:
    cudaFree(dev_ips_range_end);
}