Linking CUDA Fortran from C++

Listed below is a basic code where I am trying to use a C++ wrapper to link in some CUDA Fortran code.

/* test_c.cpp */

#include <stdio.h>
#include <iostream>

extern "C" {
extern void fill_(int *);
}

int main(int argc, char **argv)
{
    int a[100*100];
    fill_(a);
    return 0;
}



C test_f.CUF

      module kernel
      contains
!============================================================! 
      attributes(global) SUBROUTINE dGPURUN(VAL)
      
      implicit none
      INTEGER*4 VAL(:,:)
      
      CALL dGPUFILL(VAL)

      RETURN
      END
!============================================================!
      attributes(device) SUBROUTINE dGPUFILL(VAL)

      implicit none
      INTEGER*4 VAL(:,:),ROW,COL
      ROW = threadidx%x +(blockidx%x-1)*blockdim%x
      COL = threadidx%y +(blockidx%y-1)*blockdim%y
      IF(ROW.LE.100.AND.COL.LE.100) THEN
         VAL(ROW,COL) = ROW + COL
      ENDIF

      RETURN
      END

      end module kernel

!*************************************************************!
!*************************************************************!

      SUBROUTINE FILL(VAL)

      use cudafor
      use kernel
      implicit none

      INTEGER*4 VAL(100,100)
      INTEGER*4, device, dimension(100,100) :: dVAL

      type(dim3) :: grid, tBlock

      tBlock = dim3(32,32,1)
      grid = dim3(ceiling(100./tBlock%x), ceiling(100./tBlock%y), 1)

      dVAL = VAL
      CALL dGPURUN<<<grid,tBlock>>>(dVAL)
      VAL = dVAL

      RETURN
      END

When I try to run this I get the following error:

[shipmanb1@beehive TEST]$ echo $LD_LIBRARY_PATH
/opt/pgi/linux86-64/2014/cuda/6.0/lib64
[shipmanb1@beehive TEST]$ pgfortran -Mcuda -Mfixed -Mextend -c test_f.CUF
[shipmanb1@beehive TEST]$ pgc++ -c test_c.cpp
[shipmanb1@beehive TEST]$ pgc++ test_c.o test_f.o -o test.x -Mcuda -pgf90libs \
> -L/opt/pgi/linux86-64/14.7/lib -lcudafor -L/opt/pgi/linux86-64/2014/cuda/6.0/lib64 \
> -lcudart
[shipmanb1@beehive TEST]$ ./test.x 
No CUDA device code available

But if I modify things to be just Fortran then it appears to work correctly:

C test.CUF

      module kernel
      contains
!============================================================! 
      attributes(global) SUBROUTINE dGPURUN(VAL)
      
      implicit none
      INTEGER*4 VAL(:,:)
      
      CALL dGPUFILL(VAL)

      RETURN
      END
!============================================================!
      attributes(device) SUBROUTINE dGPUFILL(VAL)

      implicit none
      INTEGER*4 VAL(:,:),ROW,COL
      ROW = threadidx%x +(blockidx%x-1)*blockdim%x
      COL = threadidx%y +(blockidx%y-1)*blockdim%y
      IF(ROW.LE.100.AND.COL.LE.100) THEN
         VAL(ROW,COL) = ROW + COL
      ENDIF

      RETURN
      END

      end module kernel

!*************************************************************!
!*************************************************************!

      program test

      use cudafor
      use kernel
      implicit none

      INTEGER*4 VAL(100,100)
      INTEGER*4, device, dimension(100,100) :: dVAL

      type(dim3) :: grid, tBlock

      tBlock = dim3(32,32,1)
      grid = dim3(ceiling(100./tBlock%x), ceiling(100./tBlock%y), 1)

      dVAL = VAL
      CALL dGPURUN<<<grid,tBlock>>>(dVAL)
      VAL = dVAL

      write(*,*) 'VAL(3,3)=',VAL(3,3)

      end program



[shipmanb1@beehive TEST]$ pgfortran -Mcuda -Mfixed -Mextend test.CUF -o test.x
[shipmanb1@beehive TEST]$ ./test.x
 VAL(3,3)=            6

Is it possible to have a C++ code link in a CUDA Fortran code?

Thanks!

Hi,

I will check with our C++ expert on this and follow up soon.

Thanks,

+chris