Hello everyone,
I am new to CUDA Fortran, and I am having some issues compiling a file.
Can anyone shed some light on the problem?
The errors are below, followed by the code from the file.
Thanks!
pgfortran -c -o cudacompute.o -Mcuda -r8 -i8 -pg cudacompute.f
PGF90-S-0070-Incorrect sequence of statements (cudacompute.f: 55)
PGF90-S-0070-Incorrect sequence of statements (cudacompute.f: 66)
PGF90-S-0070-Incorrect sequence of statements (cudacompute.f: 71)
PGF90-S-0070-Incorrect sequence of statements (cudacompute.f: 174)
PGF90-S-0072-Assignment operation illegal to external procedure dslice (cudacompute.f: 224)
PGF90-S-0072-Assignment operation illegal to external procedure pslice (cudacompute.f: 225)
PGF90-S-0072-Assignment operation illegal to external procedure gravityon (cudacompute.f: 226)
PGF90-S-0072-Assignment operation illegal to external procedure grslice (cudacompute.f: 227)
PGF90-S-0072-Assignment operation illegal to external procedure geslice (cudacompute.f: 228)
PGF90-S-0072-Assignment operation illegal to external procedure uslice (cudacompute.f: 229)
PGF90-S-0072-Assignment operation illegal to external procedure vslice (cudacompute.f: 230)
PGF90-S-0072-Assignment operation illegal to external procedure wslice (cudacompute.f: 231)
PGF90-S-0072-Assignment operation illegal to external procedure cellwidthtemp_1 (cudacompute.f: 232)
PGF90-S-0072-Assignment operation illegal to external procedure flatten (cudacompute.f: 233)
PGF90-S-0072-Assignment operation illegal to external procedure is (cudacompute.f: 236)
PGF90-S-0072-Assignment operation illegal to external procedure ie (cudacompute.f: 237)
PGF90-S-0072-Assignment operation illegal to external procedure js (cudacompute.f: 238)
PGF90-S-0072-Assignment operation illegal to external procedure dualenergyformalism (cudacompute.f: 240)
PGF90-S-0072-Assignment operation illegal to external procedure dualenergyformalismeta1 (cudacompute.f: 241)
PGF90-S-0072-Assignment operation illegal to external procedure dualenergyformalismeta2 (cudacompute.f: 242)
PGF90-S-0072-Assignment operation illegal to external procedure ppmsteepeningparameter (cudacompute.f: 243)
PGF90-S-0072-Assignment operation illegal to external procedure ppmflatteningparameter (cudacompute.f: 244)
PGF90-S-0072-Assignment operation illegal to external procedure dtfixed (cudacompute.f: 245)
PGF90-S-0072-Assignment operation illegal to external procedure gamma (cudacompute.f: 246)
PGF90-S-0072-Assignment operation illegal to external procedure minimumpressure (cudacompute.f: 247)
PGF90-F-0008-Error limit exceeded (cudacompute.f: 247)
PGF90/x86-64 Linux 10.1-0: compilation aborted
make: *** [cudacompute.o] Error 1
c=================================================================
c//////////////////// SUBROUTINE CUDACOMPUTE \\\\\\\\\\\\\\\\\\c
subroutine cudacompute(
& dslice, pslice, GravityOn, grslice, geslice, uslice,
& vslice, wslice, CellWidthTemp_1, flatten,
& GridDimension_1, GridDimension_2,
& is, ie, js, je, DualEnergyFormalism,
& DualEnergyFormalismEta1, DualEnergyFormalismEta2,
& PPMSteepeningParameter, PPMFlatteningParameter,
& dtFixed, Gamma, MinimumPressure, PressureFree,
& dls, drs, pls, prs, gels, gers, uls, urs, vls, vrs,
& wls, wrs, NumberOfColours, colslice, colls, colrs,
& pbar, ubar, eslice, diffcoef, PPMDiffusionParameter,
& df, ef, uf, vf, gef, colf
& )
c
c COMBINES 3 FORTRAN ROUTINES FOR USE WITH CUDA
c
c written by: Chris Heuser
c date: July, 2010
c
c
c
c
c establish arrays on device, transfer data over, and call the kernels:
use cudafor
real size
size = GridDimension_1 * GridDimension_2
real, device, allocatable :: dslice_d(:), grslice_d(:),
& geslice_d(:),
& uslice_d(:), vslice_d(:), wslice_d(:), flatten_d(:),
& dls_d(:), drs_d(:), pls_d(:), prs_d(:), gels_d(:),
& gers_d(:),
& uls_d(:), urs_d(:), vls_d(:), vrs_d(:), wls_d(:), wrs_d(:),
& colslice_d(:), colls_d(:), colrs_d(:), colf_d(:),
& pbar_d(:),
& ubar_d(:), eslice_d(:), diffcoef_d(:), df_d(:), ef_d(:),
& uf_d(:), vf_d(:), gef_d(:), pslice_d(:)
integer, device :: PPMDiffusionParameter_d, NumberOfColours_d,
& GravityOn_d, is_d, ie_d, js_d, je_d, DualEnergyFormalism_d,
& PPMSteepeningParameter_d, PPMFlatteningParameter_d,
& PressureFree_d
real, device :: CellWidthTemp_1_d, GridDimension_1_d,
& GridDimension_2_d, DualEnergyFormalismEta1_d,
& DualEnergyFormalismEta2_d, dtFixed_d, Gamma_d,
& MinimumPressure_d
allocate(dslice_d(size))
allocate(pslice_d(size))
allocate(grslice_d(size))
allocate(geslice_d(size))
allocate(uslice_d(size))
allocate(vslice_d(size))
allocate(wslice_d(size))
allocate(flatten_d(size))
allocate(dls_d(size))
allocate(drs_d(size))
allocate(pls_d(size))
allocate(prs_d(size))
allocate(gels_d(size))
allocate(gers_d(size))
allocate(uls_d(size))
allocate(urs_d(size))
allocate(vls_d(size))
allocate(vrs_d(size))
allocate(wls_d(size))
allocate(wrs_d(size))
allocate(pbar_d(size))
allocate(ubar_d(size))
allocate(eslice_d(size))
allocate(diffcoef_d(size))
allocate(df_d(size))
allocate(ef_d(size))
allocate(uf_d(size))
allocate(vf_d(size))
allocate(ef_d(size))
allocate(gef_d(size))
allocate(colslice_d(NumberOfColours * size))
allocate(colls_d(NumberOfColours * size))
allocate(colrs_d(NumberOfColours * size))
allocate(colf_d(NumberOfColours * size))
dslice_d = dslice(size)
pslice_d = pslice(size)
GravityOn_d = GravityOn(size)
grslice_d = grslice(size)
geslice_d = geslice(size)
uslice_d = uslice(size)
vslice_d = vslice(size)
wslice_d = wslice(size)
CellWidthTemp_1_d = CellWidthTemp_1(size)
flatten_d = flatten(size)
GridDimension_1_d = GridDimension_1
GridDimension_2_d = GridDimension_2
is_d = is(size)
ie_d = ie(size)
js_d = js(size)
je_d = jd(size)
DualEnergyFormalism_d = DualEnergyFormalism(size)
DualEnergyFormalismEta1_d = DualEnergyFormalismEta1(size)
DualEnergyFormalismEta2_d = DualEnergyFormalismEta2(size)
PPMSteepeningParameter_d = PPMSteepeningParameter(size)
PPMFlatteningParameter_d = PPMFlatteningParameter(size)
dtFixed_d = dtFixed(size)
Gamma_d = Gamma(size)
MinimumPressure_d = MinimumPressure(size)
PressureFree_d = PressureFree(size)
dls_d = dls(size)
drs_d = drs(size)
pls_d = pls(size)
prs_d = prs(size)
gels_d = gels(size)
gers_d = gers(size)
uls_d = uls(size)
urs_d = urs(size)
vls_d = vls(size)
vrs_d = vrs(size)
wls_d = wls(size)
wrs_d = wrs(size)
NumberOfColours_d = NumberOfColours
colslice_d = colslice(NumberOfColours * size)
colls_d = colls(NumberOfColours * size)
colrs_d = colrs(NumberOfColours * size)
pbar_d = pbar(size)
ubar_d = ubar(size)
eslice_d = eslice(size)
diffcoef_d = diffcoef(size)
PPMDiffusionParameter_d = PPMDiffusionParameter(size)
df_d = df(size)
ef_d = ef(size)
uf_d = uf(size)
vf_d = vf(size)
ef_d = ef(size)
gef_d = gef(size)
colf_d = colf(NumberOfColours * size)
type (dim3) dimGrid, dimBlock
dimGrid= dim3(((GridDimension_1/16)+1),((GridDimension_2/16)+1),1)
dimBlock = dim3( 16, 16, 1 )
call inteuler_d<<<dimGrid,dimBlock>>>(dslice_d, pslice_d,
& GravityOn_d, grslice_d, geslice_d, uslice_d, vslice_d,
& wslice_d, CellWidthTemp_1_d, flatten_d,
& GridDimension_1_d,
& GridDimension_2_d, is_d, ie_d, js_d, je_d,
& DualEnergyFormalism_d, DualEnergyFormalismEta1_d,
& DualEnergyFormalismEta2_d, PPMSteepeningParameter_d,
& PPMFlatteningParameter_d, dtFixed_d, Gamma_d,
& PressureFree_d, dls_d, drs_d, pls_d, prs_d, gels_d,
& gers_d, uls_d, urs_d, vls_d, vrs_d, wls_d, wrs_d,
& NumberOfColours_d, colslice_d, colls_d, colrs_d
& )
call twoshock_d<<<dimGrid,dimBlock>>>(dls_d, drs_d, pls_d,
& prs_d, uls_d, urs_d, GridDimension_1_d,
& GridDimension_2_d,
& is_d, ie_p1_d, js_d, je_d, dtFixed_d, Gamma_d,
& MinimumPressure_d, PressureFree_d, pbar_d, ubar_d,
& GravityOn_d, grslice_d, DualEnergyFormalism_d,
& DualEnergyFormalismEta1_d
& )
call euler_d<<<dimGrid,dimBlock>>>(dslice_d, eslice_d,
& grslice_d, geslice_d, uslice_d, vslice_d, wslice_d,
& CellWidthTemp_1_d, diffcoef_d, GridDimension_1_d,
& GridDimension_2_d, is_d, ie_d, js_d, je_d, dtFixed_d,
& Gamma_d, PPMDiffusionParameter_d, GravityOn_d,
& DualEnergyFormalism_d, DualEnergyFormalismEta1_d,
& DualEnergyFormalismEta2_d, dls_d, drs_d, pls_d, prs_d,
& gels_d, gers_d, uls_d, urs_d, vls_d, vrs_d, wls_d, wrs_d,
& pbar_d, ubar_d, df_d, ef_d, uf_d, vf_d, wf_d, gef_d,
& NumberOfColours_d, colslice_d, colls_d, colrs_d, colf_d
& )
dslice = dslice_d
pslice = pslice_d
GravityOn = GravityOn_d
grslice = grslice_d
geslice = geslice_d
uslice = uslice_d
vslice = vslice_d
wslice = wslice_d
CellWidthTemp_1 = CellWidthTemp_1_d
flatten = flatten_d
GridDimension_1 = GridDimension_1_d
GridDimension_2 = GridDimension_2_d
is = is_d
ie = ie_d
js = js_d
je = jd_d
DualEnergyFormalism = DualEnergyFormalism_d
DualEnergyFormalismEta1 = DualEnergyFormalismEta1_d
DualEnergyFormalismEta2 = DualEnergyFormalismEta2_d
PPMSteepeningParameter = PPMSteepeningParameter_d
PPMFlatteningParameter = PPMFlatteningParameter_d
dtFixed = dtFixed_d
Gamma = Gamma_d
MinimumPressure = MinimumPressure_d
PressureFree = PressureFree_d
dls = dls_d
drs = drs_d
pls = pls_d
prs = prs_d
gels = gels_d
gers = gers_d
uls = uls_d
urs = urs_d
vls = vls_d
vrs = vrs_d
wls = wls_d
wrs = wrs_d
NumberOfColours = NumberOfColours_d
colslice = colslice_d
colls = colls_d
colrs = colrs_d
pbar = pbar_d
ubar = ubar_d
eslice = eslice_d
diffcoef = diffcoef_d
PPMDiffusionParameter = PPMDiffusionParameter_d
df = df_d
ef = ef_d
uf = uf_d
vf = vf_d
ef = ef_d
gef = gef_d
colf = colf_d
deallocate(dslice_d)
deallocate(pslice_d)
deallocate(grslice_d)
deallocate(geslice_d)
deallocate(uslice_d)
deallocate(vslice_d)
deallocate(wslice_d)
deallocate(flatten_d)
deallocate(dls_d)
deallocate(drs_d)
deallocate(pls_d)
deallocate(prs_d)
deallocate(gels_d)
deallocate(gers_d)
deallocate(uls_d)
deallocate(urs_d)
deallocate(vls_d)
deallocate(vrs_d)
deallocate(wls_d)
deallocate(wrs_d)
deallocate(pbar_d)
deallocate(ubar_d)
deallocate(eslice_d)
deallocate(diffcoef_d)
deallocate(df_d)
deallocate(ef_d)
deallocate(uf_d)
deallocate(vf_d)
deallocate(ef_d)
deallocate(gef_d)
deallocate(colslice_d)
deallocate(colls_d)
deallocate(colrs_d)
deallocate(colf_d)
return
end