Visual Studio 2012 + OptiX + Thrust = compile errors

Accelerated Computing CUDA CUDA Programming and Performance
1

Hello! I have a project where I’m trying to use Thrust to do some post-processing on data produced by OptiX. OptiX works fine by itself, but I get compile errors if I #include “thrust/device_vector.h” and try to create a device vector in my main program. The errors occur not in my code but in included code.

I ran into similar errors before and I figured out that I need to put my thrust code in a .cu file so that the nvcc compiler would be used on it. Do I need to compile my main program with nvcc as well for thrust to work?

2

For any thrust codes that target a CUDA backend, you should compile any modules that include any thrust headers with nvcc.

3

OK, I opened the properties for my main program, and under the Configuration Properties > General tab, I changed the Item Type from “C/C++ Compiler” to “CUDA C/C++”. This causes the main program to be compiled using nvcc. I ran it in that configuration without adding any thrust stuff and it works fine.

I tried adding #include “thrust/device_vector.h” and then declaring a device_vector< thrust::complex > variable without even using it. This causes a compiler error, and when I trace it by clicking on the error, it looks like the error shouldn’t happen when using nvcc. Here is where the error happens:

// ========================================================================
// X Note to the user: If you’ve found this line due to a compiler error, X
// X you need to compile your code using nvcc, rather than g++ or cl.exe X
// ========================================================================
THRUST_STATIC_ASSERT( (thrust::detail::depend_on_instantiation<RandomAccessIterator, THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC>::value) );

4

You may have a project configuration issue of some sort. These can be difficult to root out on a forum like this.

There are thrust projects included with the CUDA sample codes. Can you build one of those?

For the case where you declare a device_vector and get an error, can you paste the full console output from visual studio for the build process? All of the compile commands, and the full error output. You should be able to do this just by copy/paste from the VS output window at the bottom after the compile is complete.

5

I haven’t tried the sample project. I’m about to head home for the day but I’ll give that a shot tomorrow.

Here’s the compiler output:

1>------ Build started: Project: GRIST, Configuration: Debug x64 ------
1> Creating directories “PTX_files” if they don’t exist yet
1> Compiling CUDA source file src\GRIST.cpp…
1>
1> C:\Users\csmith\Documents\SandBox\Programs\GildedMonkey\GRIST>“C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin\nvcc.exe” --use-local-env --cl-version 2012 -ccbin “C:\Program Files (x86)\Microsoft Visual Studio 11.0\VC\bin\x86_amd64” -I"C:/ProgramData/NVIDIA Corporation/OptiX SDK 3.9.0/include" -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\include" -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\include\optixu" -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\SDK\sutil" -I.\include -I.\CUDA_files -I…....\Utils -I…....\DCS\src\wideband_utilities\signal_processing_utilities -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\include" -G --keep-dir x64\Debug --machine 64 -ptx -cudart shared -D_USE_MATH_DEFINES -DNOMINMAX -DCUDA_64_BIT_DEVICE_CODE --compiler-options /D_USE_MATH_DEFINES -DNVCC --disable-warnings -o x64\Debug\GRIST.cpp.obj “C:\Users\csmith\Documents\SandBox\Programs\GildedMonkey\GRIST\src\GRIST.cpp”
1> GRIST.cpp
1>C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/system/cuda/detail/for_each.inl(109): error C2027: use of undefined type ‘thrust::detail::STATIC_ASSERTION_FAILURE’
1> with
1> [
1> x=false
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/for_each.inl(71) : see reference to function template instantiation ‘RandomAccessIterator thrust::system::cuda::detail::for_each_nthrust::system::cuda::detail::tag,InputIterator,Size,UnaryFunction(thrust::system::cuda::detail::execution_policythrust::system::cuda::detail::tag &,RandomAccessIterator,Size,UnaryFunction)’ being compiled
1> with
1> [
1> RandomAccessIterator=thrust::device_ptr<thrust::complex>,
1> InputIterator=thrust::device_ptr<thrust::complex>,
1> Size=__int64,
1> UnaryFunction=thrust::detail::allocator_traits_detail::gozer
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/allocator/destroy_range.inl(136) : see reference to function template instantiation ‘InputIterator thrust::for_each_n<DerivedPolicy,Pointer,Size,thrust::detail::allocator_traits_detail::gozer>(const thrust::detail::execution_policy_base &,InputIterator,Size,UnaryFunction)’ being compiled
1> with
1> [
1> InputIterator=thrust::device_ptr<thrust::complex>,
1> DerivedPolicy=thrust::system::cuda::detail::tag,
1> Pointer=thrust::device_ptr<thrust::complex>,
1> Size=__int64,
1> UnaryFunction=thrust::detail::allocator_traits_detail::gozer
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/allocator/destroy_range.inl(157) : see reference to function template instantiation ‘void thrust::detail::allocator_traits_detail::destroy_range<Allocator,Pointer,Size>(Allocator &,Pointer,Size)’ being compiled
1> with
1> [
1> Allocator=thrust::device_malloc_allocator<thrust::complex>,
1> Pointer=thrust::device_ptr<thrust::complex>,
1> Size=__int64
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/contiguous_storage.inl(256) : see reference to function template instantiation ‘void thrust::detail::destroy_range<thrust::device_malloc_allocator,Base,__int64>(Allocator &,Pointer,Size)’ being compiled
1> with
1> [
1> T=thrust::complex,
1> Base=thrust::device_ptr<thrust::complex>,
1> Allocator=thrust::device_malloc_allocator<thrust::complex>,
1> Pointer=thrust::device_ptr<thrust::complex>,
1> Size=__int64
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/contiguous_storage.inl(255) : while compiling class template member function ‘void thrust::detail::contiguous_storage<T,Alloc>::destroy(thrust::detail::normal_iterator,thrust::detail::normal_iterator)’
1> with
1> [
1> T=thrust::complex,
1> Alloc=thrust::device_malloc_allocator<thrust::complex>,
1> Pointer=thrust::device_ptr<thrust::complex>
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/vector_base.inl(474) : see reference to function template instantiation ‘void thrust::detail::contiguous_storage<T,Alloc>::destroy(thrust::detail::normal_iterator,thrust::detail::normal_iterator)’ being compiled
1> with
1> [
1> T=thrust::complex,
1> Alloc=thrust::device_malloc_allocator<thrust::complex>,
1> Pointer=thrust::device_ptr<thrust::complex>
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/detail/vector_base.inl(44) : while compiling class template member function ‘thrust::detail::vector_base<T,Alloc>::vector_base(void)’
1> with
1> [
1> T=thrust::complex,
1> Alloc=thrust::device_malloc_allocator<thrust::complex>
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/device_vector.h(69) : see reference to function template instantiation ‘thrust::detail::vector_base<T,Alloc>::vector_base(void)’ being compiled
1> with
1> [
1> T=thrust::complex,
1> Alloc=thrust::device_malloc_allocator<thrust::complex>
1> ]
1> C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v7.5/include\thrust/device_vector.h(54) : see reference to class template instantiation ‘thrust::detail::vector_base<T,Alloc>’ being compiled
1> with
1> [
1> T=thrust::complex,
1> Alloc=thrust::device_malloc_allocator<thrust::complex>
1> ]
1> C:/Users/csmith/Documents/SandBox/Programs/GildedMonkey/GRIST/src/GRIST.cpp(120) : see reference to class template instantiation ‘thrust::device_vector’ being compiled
1> with
1> [
1> T=thrust::complex
1> ]
1>C:\Program Files (x86)\MSBuild\Microsoft.Cpp\v4.0\V110\BuildCustomizations\CUDA 7.5.targets(604,9): error MSB3721: The command ““C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin\nvcc.exe” --use-local-env --cl-version 2012 -ccbin “C:\Program Files (x86)\Microsoft Visual Studio 11.0\VC\bin\x86_amd64” -I"C:/ProgramData/NVIDIA Corporation/OptiX SDK 3.9.0/include” -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\include" -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\include\optixu" -I"C:\ProgramData\NVIDIA Corporation\OptiX SDK 3.9.0\SDK\sutil" -I.\include -I.\CUDA_files -I…....\Utils -I…....\DCS\src\wideband_utilities\signal_processing_utilities -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\include" -G --keep-dir x64\Debug --machine 64 -ptx -cudart shared -D_USE_MATH_DEFINES -DNOMINMAX -DCUDA_64_BIT_DEVICE_CODE --compiler-options /D_USE_MATH_DEFINES -DNVCC --disable-warnings -o x64\Debug\GRIST.cpp.obj “C:\Users\csmith\Documents\SandBox\Programs\GildedMonkey\GRIST\src\GRIST.cpp”" exited with code 2.
========== Build: 0 succeeded, 1 failed, 0 up-to-date, 0 skipped ==========

6

cuda code (including thrust code that targets the CUDA backend) should be in .cu files, not in .cpp files.

By default, when you give a .cpp file to nvcc to compile, it simply hands it off to the host compiler.

Rename your GRIST.cpp to GRIST.cu

Alternatively, you can hunt around in the VS CUDA project configuration to see if you can get it to specify the -x cu switch when compiling your GRIST.cpp with nvcc. That will override the nvcc default behavior, and force it to treat the .cpp file as if it were a .cu file.

I don’t happen to remember if or where this switch is implemented in VS CUDA projects, but I think it is in there somewhere. It may be just easier to rename your source file to GRIST.cu

7

Thanks for your help, txbob!

So I renamed my GRIST.cpp to GRIST.cu and now I get errors in sutil.h about RTcontext not being defined. I think it might be because in optix.h, when CUDACC is defined, it includes the device code instead of the host code. Maybe the solution is to give up on Thrust and write the kernels I need myself, and call them using cuLaunchKernel, etc. That has worked in other projects my colleagues have done. I was hoping to use Thrust though, since I assume it has the best algorithm for doing sum-reduce.

8

Presumably it would not be hard to partition your project into separate modules, one containing your optix code and one containing your thrust code. Use wrapper functions or the like to connect the two modules.

Your optix module can be compiled as you normally would. The thrust module should be in a .cu file and compiled with nvcc.

You can link these two modules together easily enough in a VS project.

9

Yes, that worked! I just had to use the samples as a reference to see how it should be done. In the end I left my main program alone (i.e. reverted to using the default compiler) and was able to put all my thrust stuff into .cu files compiled separately, then called from the main program. Many thanks, txbob!