Re: [eigen] Using Eigen in CUDA kernels

[Benjamin Schindler <bschindler@xxxxxxxxxxx>]

Hi

The problem is/was, that nvcc was not good enough at C++ to be able to
parse eigen. As a result, one was not able to use eigen-classes in *.cu
files, and had to resort to a separate fixed-size linear algebra
library, duplicating efforts

Regards
Benjamin


On 02/15/2013 12:45 PM, Adolfo Rodríguez Tsouroukdissian wrote:
> 
> 
> On Thu, Feb 7, 2013 at 7:21 PM, Gael Guennebaud
> <gael.guennebaud@xxxxxxxxx <mailto:gael.guennebaud@xxxxxxxxx>> wrote:
> 
>     Hi list,
> 
>     good news, NVCC (the CUDA compiler from NVidia) version 5 is now able
>     to parse and compile Eigen :) That means using Eigen in our cuda
>     kernels is "just" a matter of qualifying all our functions callable
>     from CUDA by __device__ when compiled with nvcc. As a proof of concept
>     I've created a fork there:
> 
>     https://bitbucket.org/ggael/eigen-nvcc
> 
>     This port is far to be complete, but I've been able to do stupid
>     stuff like:
> 
>     Matrix3f M(data1+tid*9);
>     Vector3f v = M.inverse() * Map<Vector3f>(data+tid*3);
>     output[tid] = v.cwiseAbs().maxCoeff();
> 
>     where data1, data, and output are device buffers. The generated
>     assembly looked reasonable.
> 
>     So no more crappy home-made fixed-size matrix and vector classes!
> 
> 
> Is this statement referring to Eigen's existing fixed-size matrices and
> vectors?. If so, I don't get it.
> 
> Adolfo.
>  
> 
> 
>     This also means that evaluating very large *coefficient-wise
>     expressions* on CUDA should not be very difficult. We just need a
>     special DeviceMatrix storing its data in device memory, and specialize
>     its assignment operator to call a CUDA kernel templated with the
>     destination and source expression...
> 
> 
>     have fun,
>     Gael.
> 
> 
> 
> 
> 
> -- 
> Adolfo Rodríguez Tsouroukdissian
> Senior robotics engineer
> adolfo.rodriguez@xxxxxxxxxxxxxxxx <mailto:adolfo.rodriguez@xxxxxxxxxxxxxxxx>
> http://www.pal-robotics.com <http://www.pal-robotics.com/>
> 
> PAL ROBOTICS S.L
> c/ Pujades 77-79, 4º4ª
> 08005 Barcelona, Spain.
> Tel. +34.93.414.53.47
> Fax.+34.93.209.11.09
> Skype: adolfo.pal-robotics
> Facebook <http://www.facebook.com/palrobotics1> - Twitter
> <http://twitter.com/#%21/palrobotics> - PAL Robotics YouTube Channel
> <http://www.youtube.com/user/PALRobotics>
> 
> AVISO DE CONFIDENCIALIDAD: Este mensaje y sus documentos adjuntos,
> pueden contener información privilegiada y/o confidencial que está
> dirigida exclusivamente a su destinatario. Si usted recibe este mensaje
> y no es el destinatario indicado, o el empleado encargado de su entrega
> a dicha persona, por favor, notifíquelo inmediatamente y remita el
> mensaje original a la dirección de correo electrónico indicada.
> Cualquier copia, uso o distribución no autorizados de esta comunicación
> queda estrictamente prohibida.
> 
> CONFIDENTIALITY NOTICE: This e-mail and the accompanying document(s) may
> contain confidential information which is privileged and intended only
> for the individual or entity to whom they are addressed.  If you are not
> the intended recipient, you are hereby notified that any disclosure,
> copying, distribution or use of this e-mail and/or accompanying
> document(s) is strictly prohibited.  If you have received this e-mail in
> error, please immediately notify the sender at the above e-mail address.