title: error : identifier "atomicOr" is undefined in Thrust program
body: <p>I have found that the Cuda atomicOr function is not recognized in my Thrust program compiled in Visual Studio 2012.</p>
  
  <p>I have read that all header files should already be included when the NVidia nvcc compiler is invoked.  Most postings on this issue state that this must mean the architectural settings are incorrect.</p>
  
  <p>I have tried it with these settings based on other postings:
  <a href="https://stackoverflow.com/questions/14411435/how-to-set-cuda-compiler-flags-in-visual-studio-2010">How to set CUDA compiler flags in Visual Studio 2010?</a></p>
  
  <p>...as well as using:
  <a href="http://s1240.photobucket.com/user/fireshot8888/media/cuda_settings.png.html" rel="nofollow noreferrer">http://s1240.photobucket.com/user/fireshot8888/media/cuda_settings.png.html</a></p>
  
  <p>main.cpp:</p>
  
  <pre><code>#include &lt;thrust/device_vector.h&gt;
  #include &lt;cstdlib&gt;
  #include &lt;iostream&gt;
  
  #include "cuda.h"
  
  using namespace std;
  
  //Visual C++ compiled main function to launch the GPU calling code
  int main(int argc, char *argv[])
  {
      //Just some random data hand keyed to make it a complete example for stack overflow while not being too complicated
      float data[] = {1.2, 3.4, 3.4, 3.3, 4.4, 4.4, 4.4, 3.4, 4.4, 4.4,
      1.2, 3.4, 3.4, 3.3, 4.4, 4.4, 4.4, 3.4, 4.4, 4.4};
  
      thrust::host_vector&lt;float&gt; h_data(data, data+20);   //Holds the contents of the file as they are read; it will be cleared once we are done with it.
  
      const int numVars = 10;
      int numBins = 4;
      int rowCount = 2;
  
      doHistogramGPU(numVars, h_data, numBins, rowCount);
  
  
      return 0;
  
  }
  </code></pre>
  
  <p>cuda.cu:</p>
  
  <pre><code>#include "cuda.h"
  
  #include &lt;iostream&gt;
  
  #include &lt;thrust/device_vector.h&gt;
  #include &lt;thrust/iterator/constant_iterator.h&gt;
  
  //I GAVE THIS A TRY BUT IT DID NOT FIX MY ISSUE::::
  #include &lt;cuda_runtime.h&gt;
  #include &lt;cuda.h&gt;
  
  using namespace std;
  
  //Function to call the kernel
  void doHistogramGPU(int numVars, thrust::host_vector&lt;float&gt; h_buffer, int numBins, int numRecords)
  {
      int dataSize = sizeof(BYTE_UNIT);
      int shiftSize = dataSize - 1;
  
      thrust::device_vector&lt;float&gt; d_buffer(h_buffer.begin(), h_buffer.end());
  
      int bitVectorSize = ceil(numRecords * numVars / (float)dataSize);
  
      thrust::device_vector&lt;BYTE_UNIT&gt; d_bitData(bitVectorSize * numBins);
  
      thrust::counting_iterator&lt;int&gt; counter(0);
      auto zipInFirst = thrust::make_zip_iterator(thrust::make_tuple(d_buffer.begin(), counter));
      auto zipInLast = thrust::make_zip_iterator(thrust::make_tuple(d_buffer.end(), counter + d_buffer.size()));
  
  
      float minValues[] = {579.8, 72.16, 0.000385, 7.576e-005, 6.954e-005, 0, 0, 2.602e-012, 1.946e-013, 7.393e-015};
      float maxValues[] = {1053, 22150, 0.7599, 0.7596, 0.24, 0.2398, 0.1623, 1.167e-007, 4.518e-006, 5.322e-008};
  
      //Get things loaded onto the device then call the kernel
      thrust::device_vector&lt;float&gt; d_minValues(minValues, minValues+10);
      thrust::device_vector&lt;float&gt; d_maxValues(maxValues, maxValues+10);
  
      thrust::device_ptr&lt;float&gt; minDevPtr = &amp;d_minValues[0];
      thrust::device_ptr&lt;float&gt; maxDevPtr = &amp;d_maxValues[0];
      thrust::device_ptr&lt;BYTE_UNIT&gt; dataDevPtr = &amp;d_bitData[0];
  
      //Invoke the Thrust Kernel
      thrust::for_each(zipInFirst, zipInLast, BinFinder(thrust::raw_pointer_cast(dataDevPtr), thrust::raw_pointer_cast(minDevPtr), thrust::raw_pointer_cast(maxDevPtr), numVars, numBins, numRecords));
  
      cout &lt;&lt; endl;
  
      return;
  
  
  
  }
  </code></pre>
  
  <p>cuda.h:</p>
  
  <pre><code>#ifndef CUDA_H
  #define CUDA_H
  
  #include &lt;thrust/device_vector.h&gt;
  #include &lt;iostream&gt;
  
  //I tried these here, too...
  #include &lt;cuda_runtime.h&gt;
  #include &lt;cuda.h&gt;
  
  
  using namespace std;
  
  
  typedef long BYTE_UNIT; //32 bit storage
  
  void doHistogramGPU(int numvars, thrust::host_vector&lt;float&gt; h_buffer, int numBins, int numRecords);
  
  struct BinFinder
  {
      BYTE_UNIT * data;
      float * rawMinVector;
      float * rawMaxVector;
      int numVars;
      int numBins;
      int numRecords;
  
  
      BinFinder(BYTE_UNIT * data, float * rawMinVector, float * rawMaxVector, int numVars, int numBins, int numRecords)
      {
          this -&gt; data = data;
          this -&gt; rawMinVector = rawMinVector;
          this -&gt; rawMaxVector = rawMaxVector;
          this -&gt; numVars = numVars;
          this -&gt; numBins = numBins;
          this -&gt; numRecords = numRecords;
      }
  
      //This kernel converts the multidimensional bin representation to a single dimensional representation
      template &lt;typename Tuple&gt;
      __device__ void operator()( Tuple param ) 
      {
          int dataSize = sizeof(BYTE_UNIT);
          int shiftSize = dataSize - 1;
  
          int bitVectorSize = ceil(numRecords * numVars / float(dataSize));
  
  
          float value = thrust::get&lt;0&gt;(param);
          int id = thrust::get&lt;1&gt;(param); 
  
          //Look up the min and max values for this data column using the index
          float min = rawMinVector[id % numVars];
          float max = rawMaxVector[id % numVars];
  
  
          //Calculate the bin id
          float percentage = (value - min) / float(max - min);
  
          char bin = percentage * numBins;
  
          if (bin == numBins)
          {
              bin--;
          }
  
          //////////////////////////////////////////////////////////////
          //Set a 1 in the appropriate bitvector for the calculated bin
          //////////////////////////////////////////////////////////////      
  
          //What I originally tried to do that appeared to have generated race conditions (using data from a file):
          //data[bin * bitVectorSize + id / dataSize] |= (1 &lt;&lt; (shiftSize - id % dataSize));   
  
          //What I've been trying to do now that generates a compilation error:
          atomicOr(data + (bin * bitVectorSize + id / dataSize), 1 &lt;&lt; (shiftSize - id % dataSize)); //&lt;----THIS DOESN'T COMPILE!!!!!!!!!
  
  }
  
  };
  
  
  #endif
  </code></pre>
  
  <p>nvcc command for cuda.cu (which includes my cuda.h file):</p>
  
  <pre><code> "C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v6.0/bin/nvcc.exe" "C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu" -c -o "C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/build/CMakeFiles/CudaLib.dir//Debug/CudaLib_generated_cuda.cu.obj" -ccbin "C:/Program Files (x86)/Microsoft Visual Studio 11.0/VC/bin" -m64 -Xcompiler ,\"/DWIN32\",\"/D_WINDOWS\",\"/W3\",\"/GR\",\"/EHsc\",\"/D_DEBUG\",\"/MDd\",\"/Zi\",\"/Ob0\",\"/Od\",\"/RTC1\" -DNVCC "-IC:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v6.0/include" "-IC:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v6.0/include"
  </code></pre>
  
  <p>Full error output by nvcc:</p>
  
  <p>1>nvcc : warning : The 'compute_10' and 'sm_10' architectures are deprecated, and may be removed in a future release.
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(107): warning : variable "minValues" was declared but never referenced
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(108): warning : variable "maxValues" was declared but never referenced
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(462): warning : variable "shiftSize" was declared but never referenced
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(602): warning : initial value of reference to non-const must be an lvalue
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(618): warning : dynamic initialization in unreachable code
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(522): warning : variable "shiftSize" was declared but never referenced
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(975): warning : initial value of reference to non-const must be an lvalue
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(993): warning : initial value of reference to non-const must be an lvalue
  1><br>
  1>C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(1022): warning : variable "shiftSize" was declared but never referenced
  1><br>
  1>c:\users\datahead8888\documents\visual studio 2012\projects\thrust-space-data\src\cuda.h(188): error : identifier "atomicOr" is undefined
  1>            detected during:
  1>              instantiation of "void BinFinder::operator()(Tuple) [with Tuple=thrust::detail::tuple_of_iterator_references]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/detail/function.h(119): here
  1>              instantiation of "Result thrust::detail::device_function::operator()(const Argument &amp;) const [with Function=BinFinder, Result=void, Argument=thrust::detail::tuple_of_iterator_references, int, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/for_each.inl(82): here
  1>              instantiation of "thrust::system::cuda::detail::for_each_n_detail::for_each_n_closure::result_type thrust::system::cuda::detail::for_each_n_detail::for_each_n_closure::operator()() [with RandomAccessIterator=thrust::zip_iterator>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, Size=unsigned int, UnaryFunction=BinFinder, Context=thrust::system::cuda::detail::detail::blocked_thread_array]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/detail/launch_closure.inl(49): here
  1>              instantiation of "void thrust::system::cuda::detail::detail::launch_closure_by_value(Closure) [with Closure=thrust::system::cuda::detail::for_each_n_detail::for_each_n_closure>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, unsigned int, BinFinder, thrust::system::cuda::detail::detail::blocked_thread_array>]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/detail/launch_closure.inl(77): here
  1>              instantiation of "thrust::system::cuda::detail::detail::closure_launcher_base::launch_function_t thrust::system::cuda::detail::detail::closure_launcher_base::get_launch_function() [with Closure=thrust::system::cuda::detail::for_each_n_detail::for_each_n_closure>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, unsigned int, BinFinder, thrust::system::cuda::detail::detail::blocked_thread_array>, launch_by_value=true]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/detail/launch_closure.inl(185): here
  1>              [ 2 instantiation contexts not shown ]
  1>              instantiation of "thrust::tuple thrust::system::cuda::detail::for_each_n_detail::configure_launch(Size) [with Closure=thrust::system::cuda::detail::for_each_n_detail::for_each_n_closure>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, unsigned int, BinFinder, thrust::system::cuda::detail::detail::blocked_thread_array>, Size=long long]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/for_each.inl(163): here
  1>              instantiation of "RandomAccessIterator thrust::system::cuda::detail::for_each_n(thrust::system::cuda::detail::execution_policy &amp;, RandomAccessIterator, Size, UnaryFunction) [with DerivedPolicy=thrust::system::cuda::detail::tag, RandomAccessIterator=thrust::zip_iterator>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, Size=long long, UnaryFunction=BinFinder]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/system/cuda/detail/for_each.inl(191): here
  1>              instantiation of "RandomAccessIterator thrust::system::cuda::detail::for_each(thrust::system::cuda::detail::execution_policy &amp;, RandomAccessIterator, RandomAccessIterator, UnaryFunction) [with DerivedPolicy=thrust::system::cuda::detail::tag, RandomAccessIterator=thrust::zip_iterator>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, UnaryFunction=BinFinder]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/detail/for_each.inl(43): here
  1>              instantiation of "InputIterator thrust::for_each(const thrust::detail::execution_policy_base &amp;, InputIterator, InputIterator, UnaryFunction) [with DerivedPolicy=thrust::system::cuda::detail::tag, InputIterator=thrust::zip_iterator>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, UnaryFunction=BinFinder]" 
  1>  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v6.0\include\thrust/detail/for_each.inl(57): here
  1>              instantiation of "InputIterator thrust::for_each(InputIterator, InputIterator, UnaryFunction) [with InputIterator=thrust::zip_iterator>, thrust::counting_iterator, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type, thrust::null_type>>, UnaryFunction=BinFinder]" 
  1>  C:/Users/datahead8888/Documents/Visual Studio 2012/Projects/thrust-space-data/src/cuda.cu(597): here
  1><br>
  1>  1 error detected in the compilation of "C:/Users/DATAHE~1/AppData/Local/Temp/tmpxft_00001f78_00000000-8_cuda.cpp1.ii".
  1>  cuda.cu</p>