Thrust: multiple definition of b40c_thrust::vote_reduction

Question

I'm using Cuda 5.0 in separate compilation mode. In

thrust/system/cuda/detail/detail/b40c/kernel_utils.h

there is this definition

__shared__ int vote_reduction[B40C_WARP_THREADS];

The linker complains of multiple definition of vote_reduction.

What is the workaround for this?

Added: Code to reproduce the problem

Thrust version: 100600

iterator.h

#pragma once
#include <thrust/transform_reduce.h>
#include <thrust/functional.h>

struct Unary_Op
{
    __host__ __device__ int operator()(const int index) const;
};

int iterates(int start, int end);

iterator.cu

#include "iterator.h"

__host__ __device__ int Unary_Op::operator()(const int index) const
{
    return index;
}

int iterates(int start, int end)
{
    thrust::counting_iterator<int> first(start);
    thrust::counting_iterator<int> last = first + end;

    Unary_Op unary_op = Unary_Op();
    thrust::plus<int> binary_op;
    int init = 0;

    int sum = thrust::transform_reduce(first, last, unary_op, init, binary_op);

    return sum;
}

calculation.h

#include "iterator.h"

int compute();

calculation.cu

#include "calculation.h"

int compute()
{
    return iterates(0,10);
}

main.cu

#include "calculation.h"

int main()
{
    compute();
    return 0;
}

compilation command (NSight)

Building file: ../calculation.cu
Invoking: NVCC Compiler
nvcc -G -g -O0 -gencode arch=compute_20,code=sm_20 -odir "" -M -o "calculation.d" "../calculation.cu"
nvcc --device-c -G -O0 -g -gencode arch=compute_20,code=sm_20  -x cu -o  "calculation.o" "../calculation.cu"

Building file: ../iterator.cu
Invoking: NVCC Compiler
nvcc -G -g -O0 -gencode arch=compute_20,code=sm_20 -odir "" -M -o "iterator.d" "../iterator.cu"
nvcc --device-c -G -O0 -g -gencode arch=compute_20,code=sm_20  -x cu -o  "iterator.o" "../iterator.cu"

Building file: ../main.cu
Invoking: NVCC Compiler
nvcc -G -g -O0 -gencode arch=compute_20,code=sm_20 -odir "" -M -o "main.d" "../main.cu"
nvcc --device-c -G -O0 -g -gencode arch=compute_20,code=sm_20  -x cu -o  "main.o" "../main.cu"

Invoking: NVCC Linker
nvcc --relocatable-device-code=true -gencode arch=compute_20,code=sm_20 -link -o  "testt"  ./calculation.o ./iterator.o ./main.o   
nvlink error   : Multiple definitions of '_ZN6thrust6system4cuda6detail6detail11b40c_thrust14vote_reductionE'
nvlink error   : Multiple definitions of '_ZN6thrust6system4cuda6detail6detail11b40c_thrust14vote_reductionE'
make: *** [tt] Error 255

Answer 1

This appeared to be a bug or issue with separate compilation that was specific to the version of thrust that shipped in the CUDA 5.x toolkit. Upgrading or downgrading to thrust 1.5.3 or thrust 1.7 appears to have solved the problem

[This answer has been assembled from comments and added as a community wiki entry to get the question off the unanswered question list]