Поддержка тяги векторных типов Cuda

В настоящее время я пытаюсь использовать функцию тяги::upper_bound. У меня возникла проблема с аргументами, которые я передаю функции. Я хотел бы использовать векторные типы CUDA, в частности double3, но когда я использую этот тип, я получаю несколько ошибок библиотеки тяги.

Блок кода, который я запускаю, приведен ниже:

/********************************************************************************                                                                            
eos_search_gpu()                                                                                                                                           
purpose        --- kernel to find the upper bound index for the                                                                                            
                 interpolation values                                                                                                                    
arguments --                                                                                                                                               

y              --- input   double3 values for which we are searching                                                                                       
my             --- input   int number of values for which we are searching                                                                                 
x              --- input   double3 array of structs containin the data table                                                                               
                         values for x, y, and f corresponding to structs                                                                                 
                         ".x", ".y", and ".z"                                                                                                            
n              --- input   int number of data values in the table                                                                                          
dim_x          --- input   int number of data values in the x-direcion of table                                                                            
j[]            --- input/output    int[]  array of int'sthat contains                                                                                      
                 the index of the (x,y,f) position of the upper bound                                                                                    


library calls --                                                                                                                                           

  __host__ __device__ ForwardIterator  thrust::upper_bound(                                                                                                  
         const thrust::detail::execution_policy_base<DerivedPolicy>& exec,                                                                               
         ForwardIterator                                             first,                                                                              
         ForwardIterator                                             last,                                                                               
         const LessThanComparable &                                  value                                                                               
         )                                                                                                                                               

 exec         --- the execution policy to use for parallelization                                                                                        
 first        --- the beginning of the ordered sequence                                                                                                  
 last         --- the end of the ordered sequence                                                                                                        
 value        --- the value to be searched.                                                                                                              

 Returns:     the furthermost iterator i, such that value < *i is false                                                                                  


 const detail::seq_t thrust::seq                                                                                                                            
 an execution policy which requires analgorithm invocation to execute                                                                                    
 sequentially in the current thread.                                                                                                                     

 ********************************************************************************/

__global__ void eos_search_gpu(const double3* y, const int my,
                           const double3* x, const int n,
                           const int dim_x, int * j){

    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if ( i < my) {
      const double ptr = thrust::upper_bound(thrust::seq, x[0].y , x[n-1].y, y[i].y);                                                                     
      j[i] = (ptr - x[i].y - 1);

    }
}

Отображаемые сообщения об ошибках следующие:

/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/iterator_traits.h(45): error: a class or namespace qualified name is required
      detected during:
        instantiation of class "thrust::iterator_traits<T> [with T=double]" 
 /opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/detail/iterator_traits.inl(53): here
        instantiation of class "thrust::iterator_difference<Iterator> [with Iterator=double]" 
  /opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/sequential/binary_search.h(102): here
        instantiation of "ForwardIterator     thrust::system::detail::sequential::upper_bound(thrust::system::detail::sequential::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
 /opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(83): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/generic/binary_search.inl(225): here
        instantiation of "ForwardIterator thrust::system::detail::generic::upper_bound(thrust::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(69): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const LessThanComparable &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, LessThanComparable=double]" 
Interpolation_cuda.cu(254): here

/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/iterator_traits.h(45): error: global-scope qualifier (leading "::") is not allowed
      detected during:
        instantiation of class "thrust::iterator_traits<T> [with T=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/detail/iterator_traits.inl(53): here
        instantiation of class "thrust::iterator_difference<Iterator> [with Iterator=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/sequential/binary_search.h(102): here
        instantiation of "ForwardIterator thrust::system::detail::sequential::upper_bound(thrust::system::detail::sequential::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
 /opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(83): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/generic/binary_search.inl(225): here
        instantiation of "ForwardIterator thrust::system::detail::generic::upper_bound(thrust::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(69): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const LessThanComparable &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, LessThanComparable=double]" 
Interpolation_cuda.cu(254): here

/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/iterator_traits.h(45): error: expected a ";"
      detected during:
        instantiation of class "thrust::iterator_traits<T> [with T=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/iterator/detail/iterator_traits.inl(53): here
        instantiation of class "thrust::iterator_difference<Iterator> [with Iterator=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/sequential/binary_search.h(102): here
        instantiation of "ForwardIterator thrust::system::detail::sequential::upper_bound(thrust::system::detail::sequential::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(83): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &, StrictWeakOrdering) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double, StrictWeakOrdering=thrust::system::detail::generic::detail::binary_search_less]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/system/detail/generic/binary_search.inl(225): here
        instantiation of "ForwardIterator thrust::system::detail::generic::upper_bound(thrust::execution_policy<DerivedPolicy> &, ForwardIterator, ForwardIterator, const T &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, T=double]" 
/opt/cudatoolkit/9.1/bin/../targets/x86_64-linux/include/thrust/detail/binary_search.inl(69): here
        instantiation of "ForwardIterator thrust::upper_bound(const thrust::detail::execution_policy_base<DerivedPolicy> &, ForwardIterator, ForwardIterator, const LessThanComparable &) [with DerivedPolicy=thrust::detail::seq_t, ForwardIterator=double, LessThanComparable=double]" 
Interpolation_cuda.cu(254): here

Мне интересно, поддерживает ли тяга использование векторных типов CUDA или я просто делаю что-то неправильно.

compiler-errors cuda thrust

Patrick Payne 10.07.2018 источник

Ответы (1)

arrow_upward
1
arrow_downward

Вам необходимо удовлетворить все ожидаемые типы входных данных для алгоритма тяги. Вы этого не делаете, так как почти каждая определенная вами величина не соответствует ожидаемой тяге.

Для начала нам понадобятся настоящие итераторы. В коде устройства это означает указатели. Thrust должен иметь возможность разыменовывать итератор/указатель, а затем вы должны указать, что делать с этим количеством. Для этого нам понадобится правильно определенный функтор. Вы можете прочитать краткое руководство по началу работы с Thrust, чтобы понять определение функтора. и использование. Наконец, разумный указатель/итератор здесь относится к типу double3, поэтому нам нужно будет создать почти все для работы с double3. Обратите внимание, что нам нужно выбрать версию upper_bound, которая позволяет определить наш собственный настраиваемый функтор, поэтому мы можем правильно манипулировать double3 количествами (то, что мы получаем, когда мы разыменовываем итераторы/указатели).

Это может помочь:

#include <thrust/binary_search.h>
#include <thrust/execution_policy.h>


struct my_comp_functor{
template <typename T>
__host__ __device__
  bool operator()(T &t1, T &t2) {
    return (t1.y < t2.y);}
};

__global__ void eos_search_gpu(const double3* y, const int my,
                           const double3* x, const int n,
                           const int dim_x, int * j, my_comp_functor my_comp){

    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if ( i < my) {
      const double3 *ptr = thrust::upper_bound(thrust::seq, x, x+n, y[i], my_comp);
      j[i] = (ptr[0].y - x[i].y - 1);

    }
}

int main(){

  double3 *d_y, *d_x;
  int *d_j;

  cudaMalloc(&d_y, 1024);
  cudaMalloc(&d_x, 1024);
  cudaMalloc(&d_j, 1024);
  struct my_comp_functor my_obj;
  eos_search_gpu<<<1,1>>>(d_y, 0, d_x, 0, 0, d_j, my_obj);
  cudaDeviceSynchronize();
}

(приведенный выше код компилируется без ошибок компиляции для меня на CUDA 9.2, но, очевидно, он не предназначен для того, чтобы быть функциональным/полезным)

В конце концов, мне кажется странным, что вы вставляете количество double в j[i] (целое число), но это ваш код.

Кроме того, я мог ошибиться в порядке в этом функторе, поэтому, возможно, вам может понадобиться изменить < на >.

Когда вы вызываете это ядро, обратите внимание, что я добавил параметр; вам нужно будет создать экземпляр объекта my_comp_functor в коде хоста, а затем передать его ядру в соответствующем месте.

Наконец, похоже, что вы выполняете векторизованный поиск, обратите внимание, что в Thrust доступны векторизованный поиск это может устранить необходимость в этом ядре.

Robert Crovella 10.07.2018

comment

Векторный поиск был именно тем, что мне было нужно. Спасибо за информацию о функторах, я не понял, что разрешить сравнение типов векторов - Patrick Payne; 20.07.2018

Поддержка тяги векторных типов Cuda

Ответы (1)

Похожие вопросы