Array of cudaTextureObject_t compiles on Windows, but not Linux

I’ve got a large amount of CUDA code developed and tested on Windows, and I want to compile it for Linux. So I installed CUDA 7.0 on Ubuntu and wrote a cmake script. After a short struggle with GCC’s quirks, everything seems to compile fine now, except this:

I need to pass an array (in device memory) of cudaTextureObject_t to a kernel, i. e.:

__global__ void MyKernel(cudaTextureObject_t* textures, float* output)
{
    output[threadIdx.x] = tex2D<float>(textures[blockIdx.x], 0.0f, 0.0f);
}

NVCC hangs up trying to compile this, with CICC running indefinitely until I kill it. No errors or warnings. Everything works fine if I pass just one cudaTextureObject_t, i. e.:

__global__ void MyKernel(cudaTextureObject_t texture, float* output)
{
    output[threadIdx.x] = tex2D<float>(texture, 0.0f, 0.0f);
}

The very same code causes no problems to NVCC on Windows. Does anyone have an idea why?

I don’t have any problems creating arrays of texture objects and passing them to a kernel, on linux, CUDA 7. Compile of a test code takes ~15 seconds.

Perhaps you could provide a short, complete test code that demonstrates the problem.

Thanks for looking into this! It gives me hope that the problem is just with my setup. Here is a minimal example that causes NVCC to hang up. It compiles just fine if d_textures is not an array.

Code.cu:

#include "cuda.h"
#include "cuda_runtime.h"

__global__ void MyKernel(cudaTextureObject_t* d_textures, float* d_output)
{
	d_output[threadIdx.x] = tex2D<float>(d_textures[0], 0.0f, 0.0f);
}

void MyFunction()
{
	cudaTextureObject_t* d_textures = 0;
	float* d_output = 0;

	dim3 grid = dim3(10, 1, 1);
	MyKernel <<<grid, 128>>> (d_textures, d_output);
}

CMakeLists.txt:

CMAKE_MINIMUM_REQUIRED(VERSION 2.8)
PROJECT(mylib)

FIND_PACKAGE(CUDA REQUIRED)

INCLUDE(FindCUDA)

INCLUDE_DIRECTORIES(/usr/local/cuda-7.0/include include)

LINK_DIRECTORIES(/usr/local/cuda-7.0/lib64)

FILE(GLOB_RECURSE SOURCES "*.cu")
CUDA_ADD_LIBRARY(mylib ${SOURCES})

LIST(APPEND CUDA_NVCC_FLAGS -gencode arch=compute_52,code=sm_52)

TARGET_LINK_LIBRARIES(mylib libcudart.so)

Oops, this is embarrassing. Problem solved by putting

LIST(APPEND CUDA_NVCC_FLAGS -gencode arch=compute_52,code=sm_52)

before

CUDA_ADD_LIBRARY(mylib ${SOURCES})