c++ - 从传递的 FORTRAN 数组生成 CUSP coo_matrix

Question

我正在致力于将 CUSP 求解器集成到现有的 FORTRAN 代码中。作为第一步，我只是尝试从 FORTRAN 中传入一对整数数组和一个 float (FORTRAN 中的实数 * 4)，它们将用于构造并打印 COO 格式的 CUSP 矩阵。

到目前为止，我已经能够关注这个线程并得到所有要编译和链接的东西:Unresolved references using IFORT with nvcc and CUSP

不幸的是，该程序显然将垃圾发送到 CUSP 矩阵并最终因以下错误而崩溃:

$./fort_cusp_test
 testing 1 2 3
sparse matrix <1339222572, 1339222572> with 1339222568 entries
libc++abi.dylib: terminating with uncaught exception of type thrust::system::system_error: invalid argument

Program received signal SIGABRT: Process abort signal.

Backtrace for this error:
#0  0x10ff86ff6
#1  0x10ff86593
#2  0x7fff8593ff19
Abort trap: 6

cuda和fortran源代码如下:

cusp_runner.cu

#include <stdio.h>
#include <cusp/coo_matrix.h>
#include <iostream>
#include <cusp/krylov/cg.h>
#include <cusp/print.h>

#if defined(__cplusplus)
extern "C" {
#endif

void test_coo_mat_print_(int * row_i, int * col_j, float * val_v, int n, int nnz ) {

   //wrap raw input pointers with thrust::device_ptr
   thrust::device_ptr<int> wrapped_device_I(row_i);
   thrust::device_ptr<int> wrapped_device_J(col_j);
   thrust::device_ptr<float> wrapped_device_V(val_v);

   //use array1d_view to wrap individual arrays
   typedef typename cusp::array1d_view< thrust::device_ptr<int> > DeviceIndexArrayView;
   typedef typename cusp::array1d_view< thrust::device_ptr<float> > DeviceValueArrayView;

   DeviceIndexArrayView row_indices(wrapped_device_I, wrapped_device_I + n);
   DeviceIndexArrayView column_indices(wrapped_device_J, wrapped_device_J + nnz);
   DeviceValueArrayView values(wrapped_device_V, wrapped_device_V + nnz);

   //combine array1d_views into coo_matrix_view
   typedef   cusp::coo_matrix_view<DeviceIndexArrayView,DeviceIndexArrayView,DeviceValueArrayView> DeviceView;

   //construct coo_matrix_view from array1d_views
   DeviceView A(n,n,nnz,row_indices,column_indices,values);

   cusp::print(A);
}
#if defined(__cplusplus)
}
#endif

fort_cusp_test.f90

program fort_cuda_test

   implicit none

interface
   subroutine test_coo_mat_print_(row_i,col_j,val_v,n,nnz) bind(C)
      use, intrinsic :: ISO_C_BINDING, ONLY: C_INT,C_FLOAT
      implicit none
      integer(C_INT) :: n, nnz, row_i(:), col_j(:)
      real(C_FLOAT) :: val_v(:)
   end subroutine test_coo_mat_print_
end interface

   integer*4   n
   integer*4   nnz

   integer*4, target :: rowI(9),colJ(9)
   real*4, target :: valV(9)

   integer*4, pointer ::   row_i(:)
   integer*4, pointer ::   col_j(:)
   real*4, pointer ::   val_v(:)

   n     =  3
   nnz   =  9
   rowI =  (/ 1, 1, 1, 2, 2, 2, 3, 3, 3/)
   colJ =  (/ 1, 2, 3, 1, 2, 3, 1, 2, 3/)
   valV =  (/ 1, 2, 3, 4, 5, 6, 7, 8, 9/)

   row_i => rowI
   col_j => colJ
   val_v => valV

   write(*,*) "testing 1 2 3"

   call test_coo_mat_print_(row_i,col_j,val_v,n,nnz)

end program fort_cuda_test

如果你想自己尝试，这是我的(相当不优雅的)makefile:

Test:
   nvcc -Xcompiler="-fPIC" -shared cusp_runner.cu -o cusp_runner.so -I/Developer/NVIDIA/CUDA-6.5/include/cusp
   gfortran -c fort_cusp_test.f90
   gfortran fort_cusp_test.o cusp_runner.so -L/Developer/NVIDIA/CUDA-6.5/lib -lcudart -o fort_cusp_test

clean:
   rm *.o *.so

库路径当然需要根据需要进行更改。

谁能为我指出正确的方向，让我了解如何从 Fortran 代码正确传递所需的数组？

---

删除接口(interface) block 并在 C 函数的开头添加打印语句后，我可以看到数组正在正确传递，但 n 和 nnz 导致了问题。我得到以下输出:

$ ./fort_cusp_test
 testing 1 2 3
n: 1509677596, nnz: 1509677592
     i,  row_i,  col_j,        val_v
     0,      1,      1,   1.0000e+00
     1,      1,      2,   2.0000e+00
     2,      1,      3,   3.0000e+00
     3,      2,      1,   4.0000e+00
     4,      2,      2,   5.0000e+00
     5,      2,      3,   6.0000e+00
     6,      3,      1,   7.0000e+00
     7,      3,      2,   8.0000e+00
     8,      3,      3,   9.0000e+00
     9,      0,  32727,   0.0000e+00
    ...
    etc
    ...
    Program received signal SIGSEGV: Segmentation fault - invalid memory reference.

Backtrace for this error:
#0  0x105ce7ff6
#1  0x105ce7593
#2  0x7fff8593ff19
#3  0x105c780a2
#4  0x105c42dbc
#5  0x105c42df4
Segmentation fault: 11

fort_cusp_test

    interface
       subroutine test_coo_mat_print_(row_i,col_j,val_v,n,nnz) bind(C)
          use, intrinsic :: ISO_C_BINDING, ONLY: C_INT,C_FLOAT
          implicit none
          integer(C_INT),value :: n, nnz
          integer(C_INT) :: row_i(:), col_j(:)
          real(C_FLOAT) :: val_v(:)
       end subroutine test_coo_mat_print_
    end interface

       integer*4   n
       integer*4   nnz

       integer*4, target :: rowI(9),colJ(9)
       real*4, target :: valV(9)

       integer*4, pointer ::   row_i(:)
       integer*4, pointer ::   col_j(:)
       real*4, pointer ::   val_v(:)

       n     =  3
       nnz   =  9
       rowI =  (/ 1, 1, 1, 2, 2, 2, 3, 3, 3/)
       colJ =  (/ 1, 2, 3, 1, 2, 3, 1, 2, 3/)
       valV =  (/ 1, 2, 3, 4, 5, 6, 7, 8, 9/)

       row_i => rowI
       col_j => colJ
       val_v => valV

       write(*,*) "testing 1 2 3"

       call test_coo_mat_print_(rowI,colJ,valV,n,nnz)

    end program fort_cuda_test

cusp_runner.cu

   #include <stdio.h>
    #include <cusp/coo_matrix.h>
    #include <iostream>
    // #include <cusp/krylov/cg.h>
    #include <cusp/print.h>

    #if defined(__cplusplus)
    extern "C" {
    #endif

    void test_coo_mat_print_(int * row_i, int * col_j, float * val_v, int n, int nnz ) {

       printf("n: %d, nnz: %d\n",n,nnz);

       printf("%6s, %6s, %6s, %12s \n","i","row_i","col_j","val_v");
       for(int i=0;i<n;i++) {
          printf("%6d, %6d, %6d, %12.4e\n",i,row_i[i],col_j[i],val_v[i]);
       }
       if ( false ) {
       //wrap raw input pointers with thrust::device_ptr
       thrust::device_ptr<int> wrapped_device_I(row_i);
       thrust::device_ptr<int> wrapped_device_J(col_j);
       thrust::device_ptr<float> wrapped_device_V(val_v);

       //use array1d_view to wrap individual arrays
       typedef typename cusp::array1d_view< thrust::device_ptr<int> > DeviceIndexArrayView;
       typedef typename cusp::array1d_view< thrust::device_ptr<float> > DeviceValueArrayView;

       DeviceIndexArrayView row_indices(wrapped_device_I, wrapped_device_I + n);
       DeviceIndexArrayView column_indices(wrapped_device_J, wrapped_device_J + nnz);
       DeviceValueArrayView values(wrapped_device_V, wrapped_device_V + nnz);

       //combine array1d_views into coo_matrix_view
       typedef cusp::coo_matrix_view<DeviceIndexArrayView,DeviceIndexArrayView,DeviceValueArrayView> DeviceView;

       //construct coo_matrix_view from array1d_views
       DeviceView A(n,n,nnz,row_indices,column_indices,values);

       cusp::print(A); }
    }
    #if defined(__cplusplus)
    }
    #endif

Answer 1

将参数从 Fortran 传递到 C 例程有两种方法:第一种是使用接口(interface) block (现代 Fortran 中的一种新方法)，第二种是不使用接口(interface) block (一种旧方法，即使对于Fortran77)。

首先，下面介绍第一种使用接口(interface) block 的方法。因为 C 例程需要接收 C 指针(row_i、col_j 和 val_v)，所以我们需要从 Fortran 端传递这些变量的地址。为此，我们必须在界面 block 中使用星号 (*) 而不是冒号 (:)，如下所示。 (如果我们使用冒号，那么这会告诉 Fortran 编译器发送 Fortran 指针对象的地址 [1]，这不是所需的行为。)另外，由于 C 例程中的 n 和 nnz 被声明为值(而不是指针) ，接口(interface) block 需要有这些变量的 VALUE 属性，以便 Fortran 编译器发送 n 和 nnz 的值而不是它们的地址。总而言之，在第一种方法中，C 和 Fortran 例程如下所示:

Fortran routine:
...
interface
    subroutine test_coo_mat_print_(row_i,col_j,val_v,n,nnz) bind(C)
        use, intrinsic :: ISO_C_BINDING, ONLY: C_INT,C_FLOAT
        implicit none
        integer(C_INT) :: row_i(*), col_j(*)
        real(C_FLOAT) :: val_v(*)
        integer(C_INT), value :: n, nnz     !! see note [2] below also
    end subroutine test_coo_mat_print_
end interface
...
call test_coo_mat_print_( rowI, colJ, valV, n, nnz )

C routine:
void test_coo_mat_print_ (int * row_i, int * col_j, float * val_v, int n, int nnz ) 

---

下面是关于没有接口(interface) block 的第二种方法。在这种方法中，首先将接口(interface) block 和数组指针完全删除，并将 Fortran 代码更改如下

Fortran routine:

integer  rowI( 9 ), colJ( 9 ), n, nnz     !! no TARGET attribute necessary
real     valV( 9 )

! ...set rowI etc as above...

call test_coo_mat_print ( rowI, colJ, valV, n, nnz )   !! "_" is dropped

C程序如下

void test_coo_mat_print_ ( int* row_i, int* col_j, float* val_v, int* n_, int* nnz_ )
{
    int n = *n_, nnz = *nnz_;

    printf( "%d %d \n", n, nnz );
    for( int k = 0; k < 9; k++ ) {
        printf( "%d %d %10.6f \n", row_i[ k ], col_j[ k ], val_v[ k ] );
    }

    // now go to thrust...
}

请注意，n_ 和 nnz_ 在 C 例程中被声明为指针，因为没有接口(interface) block ，Fortran 编译器总是将实际参数的地址发送到 C 例程。另请注意，在上面的 C 例程中，打印 row_i 等的内容以确保正确传递参数。如果打印的值是正确的，那么我猜问题更有可能出现在推力例程的调用中(包括如何传递大小信息，如 n 和 nnz)。

[1] 声明为“real, pointer::a(:)”的 Fortran 指针实际上表示类似于数组 View 类的东西(C++ 行话)，这与指向的实际数据不同。这里需要的是发送实际数据的地址，而不是这个数组 View 对象的地址。此外，接口(interface) block 中的星号 (a(*)) 表示假定大小的数组，这是 Fortran 中传递数组的旧方法。在这种情况下，数组的第一个元素的地址将按预期传递。

[2] 如果在 C 例程中将 n 和 nnz 声明为指针(如在第二种方法中)，则不应附加此 VALUE 属性，因为 C 例程需要的地址实际参数而不是它们的值。