Google Colaboratory で CUDA Fortran

#Google Colaboratory
我々の私生活を覗き見して金儲けをする悪徳商人 google が、乞食に無料で計算させてくださる Google Colaboratory ですが、Nvidia が無料で使わせてくださる CUDA Fortran をインストールすることで、Fortran から GPU が使える模様です。

本当は CUDA Fortran を Google Colab から直接 wget 出来ればいいのですが、やり方が分かりませんので、一度ローカルにダウンロードして、次に Google Drive に upload し、Google Colab と Google Drive を連結した後、Google Colab 側にコピーし、そこでインストールします。

##手順

1. CUDA Fortran download :
   https://www.pgroup.com/products/community.htm

2. pgilinux-2018-184-x86-64.tar.gz （現在の version)を各人の googledrive へ

3. Google Drive と Colab の連結を以下のページを参考に Colab 上で行う。Colab では GPU 付のノートブックで作業する。

!apt-get install -y -qq software-properties-common python-software-properties module-init-tools
!add-apt-repository -y ppa:alessandro-strada/ppa 2>&1 > /dev/null
!apt-get update -qq 2>&1 > /dev/null
!apt-get -y install -qq google-drive-ocamlfuse fuse
from google.colab import auth
auth.authenticate_user()
from oauth2client.client import GoogleCredentials
creds = GoogleCredentials.get_application_default()
import getpass
!google-drive-ocamlfuse -headless -id={creds.client_id} -secret={creds.client_secret} < /dev/null 2>&1 | grep URL
vcode = getpass.getpass()
!echo {vcode} | google-drive-ocamlfuse -headless -id={creds.client_id} -secret={creds.client_secret}

ここでトークンを聞かれるので、出たリンクを踏んだ先に表示される文字列をコピペする。

!mkdir -p drive
!google-drive-ocamlfuse drive

これで、Drive ディレクトリ下に Google Drive がリンクされる。

4.Google Drive から Colab 上に、pgilinux-2018-184-x86-64.tar.gz をコピーし、展開する。

!cp drive/pgilinux-2018-184-x86-64.tar.gz .
!mkdir PGI
!cd PGI
!tar zxvf ../pgilinux-2016-1710-x86_64.tar.gz

5.環境変数をセットしてインストール（手動インストールは可能だがメンドイ）。まず !bash でシェルを起動して、白枠の中に入力する。

!bash
root@60018d3878b3:/content/PGI# export PGI_SILENT="true"
root@60018d3878b3:/content/PGI# export PGI_ACCEPT_EULA="accept"
root@60018d3878b3:/content/PGI# ./install
Installing PGI version 18.4 into /opt/pgi
localrc has not changed
localrc has not changed
Installing module files for PGI linux86-64 version(s): 18.4 2018
Done
root@60018d3878b3:/content/PGI# exit

CELL を停止する。

6.PATH をセットできないので、絶対パスでコンパイラ起動する。

!/opt/pgi/linux86-64-llvm/18.4/bin/pgf90 -V

7.ソースファイル入力

以下のページの例でテスト
https://www.softek.co.jp/SPG/Pgi/TIPS/opt_cudaF.html

%%writefile sample.f90
module cudamod
   use cudafor
   implicit none

   contains
   !  kernel subprogram
   attributes(global) subroutine test1 (intdat,n)
      integer :: it, ib
      integer, value :: n
      integer, device :: intdat(n)
      !----------------------------
      it = threadidx%x
      ib = (blockidx%x-1) * 16
      intdat(it+ib) = it + ib
      !----------------------------
   end subroutine test1
end module cudamod

program cuda_device
   use cudafor
   use cudamod
   implicit none
   integer, parameter :: n=64
   integer :: int_h(n), ist
   ! デバイス側の配列データを allocatable で宣言
   integer, device, allocatable, dimension(:)  :: int_d

   int_h = 0
   allocate(int_d(n))
   
   ! カーネル（デバイスプログラム）を起動する
   call test1 <<<n/16,16>>> (int_d, n) !pass arguments
   ist = cudathreadsynchronize()  ! 同期ポイント

   ! デバイス側の配列データをホスト側に戻す
   int_h = int_d
   print *,'int_h = ',int_h
   deallocate(int_d)
end program cuda_device

8.コンパイル＆実行

!/opt/pgi/linux86-64-llvm/18.4/bin/pgf90 -Mcuda sample.f90

!./a.out

 int_h =             1            2            3            4            5 
            6            7            8            9           10           11 
           12           13           14           15           16           17 
           18           19           20           21           22           23 
           24           25           26           27           28           29 
           30           31           32           33           34           35 
           36           37           38           39           40           41 
           42           43           44           45           46           47 
           48           49           50           51           52           53 
           54           55           56           57           58           59 
           60           61           62           63           64