CUDA Fonctions mathématiques de raccourci de poussée

Question

Existe-t-il un moyen d'envelopper automatiquement une fonction mathématique CUDA dans un foncteur afin que l'on puisse appliquer thrust::transform sans avoir à écrire manuellement un foncteur? Quelque chose comme la fonctionnalité que (je comprends) std::function fournit?

thrust::placeholders ne semble pas aimer les fonctions mathématiques. std::function ne semble pas être disponible.

code Exemple:

#include <thrust/transform.h> 
#include <thrust/device_vector.h> 
#include <iostream> 
#include <functional> 
#include <math.h> 

struct myfunc{ 
    __device__ 
    double operator()(double x,double y){ 
    return hypot(x,y); 
    } 
}; 

int main(){ 

    double x0[10] = {3.,0.,1.,2.,3.,4.,5.,6.,7.,8.}; 
    double y0[10] = {4.,0.,1.,2.,3.,4.,5.,6.,7.,8.}; 

    thrust::device_vector<double> x(x0,x0+10); 
    thrust::device_vector<double> y(y0,y0+10); 
    thrust::device_vector<double> r(10); 

    for (int i=0;i<10;i++) std::cout << x0[i] <<" "; std::cout<<std::endl; 
    for (int i=0;i<10;i++) std::cout << y0[i] <<" "; std::cout<<std::endl; 

    // this works: 
    thrust::transform(x.begin(),x.end(),y.begin(),r.begin(), myfunc()); 

    // this doesn't compile: 
    using namespace thrust::placeholders; 
    thrust::transform(x.begin(),x.end(),y.begin(),r.begin(), hypot(_1,_2)); 

    // nor does this: 
    thrust::transform(x.begin(),x.end(),y.begin(),r.begin(), std::function<double(double,double)>(hypot)); 


    for (int i=0;i<10;i++) std::cout << r[i] <<" "; std::cout<<std::endl; 
} 

Answer 1

Conversion mon commentaire dans cette réponse:

Comme @JaredHoberock déjà dit, il n'y a aucun moyen automatique pour obtenir ce que vous voulez. Il y a toujours des frais généraux de syntaxe/dactylographie. Une façon de réduire ce temps d'écriture d'un foncteur séparé (comme vous l'avez fait avec my_func) est d'utiliser lambdas. Depuis CUDA 7.5 il y a une experimental device lambda feature qui vous permet de faire le folllowing:

auto h = []__device__(double x, double y){return hypot(x,y);}; 
thrust::transform(x.begin(),x.end(),y.begin(),r.begin(), h); 

Vous devez ajouter le commutateur du compilateur nvcc suivante pour compiler ceci:

nvcc --expt-extended-lambda ... 

---

Une autre approche consiste à convertir la fonction dans un foncteur utilisant les éléments suivants Wrapper:

template<typename Sig, Sig& S> 
struct Wrapper; 

template<typename R, typename... T, R(&function)(T...)> 
struct Wrapper<R(T...), function> 
{ 
    __device__ 
    R operator() (T&... a) 
    { 
     return function(a...); 
    } 
}; 

Vous pouvez ensuite l'utiliser comme ceci:

thrust::transform(x.begin(),x.end(),y.begin(),r.begin(), Wrapper<double(double,double), hypot>()); 

Answer 2

Comme le fait remarquer par M.S., une façon de réduire les frais généraux d'écrire un foncteur est d'utiliser lambda expressions. S'il vous plaît, notez que les lambdas GPU sont disponibles avec CUDA 8.0 RC (bien que toujours à un stade expérimental). Une autre possibilité est d'utiliser la technique fictif. Ci-dessous, il y a deux exemples de travail pour les deux cas mentionnés:

LAMBDA EXPRESSIONS

#include <thrust/device_vector.h> 
#include <thrust/functional.h> 
#include <thrust/transform.h> 

// Available for device operations only from CUDA 8.0 (experimental stage) 
// Compile with the flag --expt-extended-lambda 

using namespace thrust::placeholders; 

int main(void) 
{ 
    // --- Input data 
    float a = 2.0f; 
    float x[4] = { 1, 2, 3, 4 }; 
    float y[4] = { 1, 1, 1, 1 }; 

    thrust::device_vector<float> X(x, x + 4); 
    thrust::device_vector<float> Y(y, y + 4); 

    thrust::transform(X.begin(), 
         X.end(), 
         Y.begin(), 
         Y.begin(), 
         [=] __host__ __device__ (float x, float y) { return a * x + y; }  // --- Lambda expression 
        );   

    for (size_t i = 0; i < 4; i++) std::cout << a << " * " << x[i] << " + " << y[i] << " = " << Y[i] << std::endl; 

    return 0; 
} 

PLACEHOLDERS

#include <thrust/device_vector.h> 
#include <thrust/functional.h> 
#include <thrust/transform.h> 

using namespace thrust::placeholders; 

int main(void) 
{ 
    // --- Input data 
    float a = 2.0f; 
    float x[4] = { 1, 2, 3, 4 }; 
    float y[4] = { 1, 1, 1, 1 }; 

    thrust::host_vector<float> X(x, x + 4); 
    thrust::host_vector<float> Y(y, y + 4); 

    thrust::transform(X.begin(), X.end(), 
         Y.begin(),   
         Y.begin(), 
         a * _1 + _2); 

    for (size_t i = 0; i < 4; i++) std::cout << a << " * " << x[i] << " + " << y[i] << " = " << Y[i] << std::endl; 

    return 0; 
}