ó
àÆ÷Xc           @` s  d  Z  d d l m Z m Z m Z d d l Z d d l Z d d l Z d d l	 j
 j Z d d l Z e j j j s’ d d l m Z e d ƒ ‚ n  e j e k rÀ d d l m Z e d ƒ ‚ n  d d l Z d d l j Z d d l Z d „  Z d „  Z d	 „  Z d S(
   sÊ   
This file is an example of view the memory allocated by pycuda in a GpuArray
in a CudaNdarray to be able to use it in Theano.

This also serve as a test for the function: cuda_ndarray.from_gpu_pointer
i    (   t   absolute_importt   print_functiont   divisionN(   t   SkipTestsC   Pycuda not installed. We skip tests of Theano Ops with pycuda code.s%   Optional theano package cuda disabledc       	   C` sÍ   d d l  m }  |  d ƒ } | j d ƒ } t j j d ƒ j t j ƒ } t j j d ƒ j t j ƒ } t j | ƒ } | t	 j
 | ƒ t	 j | ƒ t	 j | ƒ d d d	 d ƒ| | | k j ƒ  sÉ t ‚ d
 S(   s2   Run pycuda only example to test that pycuda works.i    (   t   SourceModulesz   
__global__ void multiply_them(float *dest, float *a, float *b)
{
  const int i = threadIdx.x;
  dest[i] = a[i] * b[i];
}
t   multiply_themid   t   blocki  i   t   gridN(   i  i   i   (   i   i   (   t   pycuda.compilerR   t   get_functiont   npt   randomt   randnt   astypet   float32t
   zeros_liket   drvt   Outt   Int   allt   AssertionError(   R   t   modR   t   at   bt   dest(    (    sK   /tmp/pip-build-X4mzal/theano/theano/misc/tests/test_pycuda_theano_simple.pyt   test_pycuda_only   s    	'c       	   C` sß   d d l  m }  |  d ƒ } | j d ƒ } t j j d ƒ j t j ƒ } t j j d ƒ j t j ƒ } t j	 | ƒ } t j	 | ƒ } t j	 j
 | j ƒ } | | | | d d d	 d ƒt j | ƒ | | k j ƒ  sÛ t ‚ d
 S(   sB   Simple example with pycuda function and Theano CudaNdarray object.i    (   R   sz   
__global__ void multiply_them(float *dest, float *a, float *b)
{
  const int i = threadIdx.x;
  dest[i] = a[i] * b[i];
}
R   id   R   i  i   R   N(   i  i   i   (   i   i   (   R   R   R	   R
   R   R   R   R   t   cuda_ndarrayt   CudaNdarrayt   zerost   shapet   asarrayR   R   (   R   R   R   R   R   t   gat   gbR   (    (    sK   /tmp/pip-build-X4mzal/theano/theano/misc/tests/test_pycuda_theano_simple.pyt   test_pycuda_theano6   s    	c    
      C` s  t  j j d d ƒ }  t t j |  ƒ ƒ t j |  ƒ } t d d d ƒt t j |  ƒ ƒ t j |  ƒ | k su t ‚ t j j	 |  j
 Œ  j t j ƒ } t j | ƒ } d g } x7 |  j
 d  d  d	 … d	  D] } | j | d	 | ƒ qÉ Wt | d  d  d	 … ƒ } t d
 | ƒ | j | k s2t | j | f ƒ ‚ t |  j ƒ } t j | |  j
 | |  ƒ } t d t j |  ƒ ƒ t j |  ƒ | d k s‘t ‚ t j | ƒ d k j ƒ  s²t ‚ | j |  k sÇt ‚ | j ƒ  } t j |  ƒ | d k sòt ‚ | j |  k st ‚ ~ t j |  ƒ | d k s)t ‚ t j t j d g g g d d ƒƒ }	 | |	 7} t j | ƒ t j |  j
 ƒ k j ƒ  s‡t ‚ t j | ƒ d k j ƒ  s¨t ‚ | j
 | j
 k sÀt ‚ | j | j k sêt | j | j f ƒ ‚ t j | ƒ | k j ƒ  st ‚ | | 7} t j | ƒ | d k j ƒ  s:t ‚ ~ t d d d ƒt t j |  ƒ ƒ t j |  ƒ | k s{t ‚ d  S(   Ni   i   i   R   s0   gpuarray ref count before creating a CudaNdarrayt   endt    i   iÿÿÿÿt   stridess/   gpuarray ref count after creating a CudaNdarrayi    i   t   dtypes1   gpuarray ref count after deleting the CudaNdarray(   i   i   i   (   t   pycudat   gpuarrayR   t   printt   syst   getrefcountR   R
   R   R   R   R   R   R   R   t   appendt   tuplet   _stridest   intt   gpudatat   from_gpu_pointerR   R   t   baset   viewt   ones(
   t   yt   initial_refcountt   randt	   cuda_randR$   t   it   y_ptrt   zt   zzt	   cuda_ones(    (    sK   /tmp/pip-build-X4mzal/theano/theano/misc/tests/test_pycuda_theano_simple.pyt   test_pycuda_memory_to_theanoO   sN    !	!$!
-!*!
%(   t   __doc__t
   __future__R    R   R   R)   t   numpyR
   t   theanot   theano.sandbox.cudat   sandboxt   cudaR   t   theano.misc.pycuda_initt   misct   pycuda_initt   pycuda_availablet   nose.plugins.skipR   t   cuda_availablet   FalseR&   t   pycuda.drivert   driverR   t   pycuda.gpuarrayR   R!   R=   (    (    (    sK   /tmp/pip-build-X4mzal/theano/theano/misc/tests/test_pycuda_theano_simple.pyt   <module>   s$   		