/*---------------------------------------------------------------------------*\
  =========                 |
  \\      /  F ield         | OpenFOAM: The Open Source CFD Toolbox
   \\    /   O peration     |
    \\  /    A nd           | www.openfoam.com
     \\/     M anipulation  |
-------------------------------------------------------------------------------
    Copyright (C) 2011-2016 OpenFOAM Foundation
    Copyright (C) 2019 OpenCFD Ltd.
-------------------------------------------------------------------------------
License
    This file is part of OpenFOAM.

    OpenFOAM is free software: you can redistribute it and/or modify it
    under the terms of the GNU General Public License as published by
    the Free Software Foundation, either version 3 of the License, or
    (at your option) any later version.

    OpenFOAM is distributed in the hope that it will be useful, but WITHOUT
    ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or
    FITNESS FOR A PARTICULAR PURPOSE.  See the GNU General Public License
    for more details.

    You should have received a copy of the GNU General Public License
    along with OpenFOAM.  If not, see <http://www.gnu.org/licenses/>.

\*---------------------------------------------------------------------------*/

#include "cyclicGGAMGInterface.H"
#include "addToRunTimeSelectionTable.H"
#include "labelPairHashes.H"

// * * * * * * * * * * * * * * Static Data Members * * * * * * * * * * * * * //

namespace Foam
{
    defineTypeNameAndDebug(cyclicGGAMGInterface, 0);
    addToRunTimeSelectionTable
    (
        GGAMGInterface,
        cyclicGGAMGInterface,
        gpulduInterface
    );
    addToRunTimeSelectionTable
    (
        GGAMGInterface,
        cyclicGGAMGInterface,
        Istream
    );


    // Add under name cyclicSlip
    addNamedToRunTimeSelectionTable
    (
        GGAMGInterface,
        cyclicGGAMGInterface,
        gpulduInterface,
        cyclicSlip
    );
    addNamedToRunTimeSelectionTable
    (
        GGAMGInterface,
        cyclicGGAMGInterface,
        Istream,
        cyclicSlip
    );
}


// * * * * * * * * * * * * * * * * Constructors  * * * * * * * * * * * * * * //

Foam::cyclicGGAMGInterface::cyclicGGAMGInterface
(
    const label index,
    const gpulduInterfacePtrsList& coarseInterfaces,
    const gpulduInterface& fineInterface,
    const labelField& localRestrictAddressing,
    const labelField& neighbourRestrictAddressing,
    const label fineLevelIndex,
    const label coarseComm
)
:
    GGAMGInterface(index, coarseInterfaces),
    neighbPatchID_
    (
        refCast<const cyclicgpuLduInterface>(fineInterface).neighbPatchID()
    ),
    owner_(refCast<const cyclicgpuLduInterface>(fineInterface).owner()),
    forwardT_(refCast<const cyclicgpuLduInterface>(fineInterface).gpuForwardT()),
    reverseT_(refCast<const cyclicgpuLduInterface>(fineInterface).gpuReverseT())
{
    // From coarse face to coarse cell
    DynamicList<label> dynFaceCells(localRestrictAddressing.size());
    // From fine face to coarse face
    DynamicList<label> dynFaceRestrictAddressing
    (
        localRestrictAddressing.size()
    );

    // From coarse cell pair to coarse face
    labelPairLookup cellsToCoarseFace(2*localRestrictAddressing.size());

    forAll(localRestrictAddressing, ffi)
    {
        labelPair cellPair;

        // Do switching on master/slave indexes based on the owner/neighbour of
        // the processor index such that both sides get the same answer.
        if (owner())
        {
            // Master side
            cellPair = labelPair
            (
                localRestrictAddressing[ffi],
                neighbourRestrictAddressing[ffi]
            );
        }
        else
        {
            // Slave side
            cellPair = labelPair
            (
                neighbourRestrictAddressing[ffi],
                localRestrictAddressing[ffi]
            );
        }

        const auto fnd = cellsToCoarseFace.cfind(cellPair);

        if (fnd.found())
        {
            // Already have coarse face
            dynFaceRestrictAddressing.append(fnd.val());
        }
        else
        {
            // New coarse face
            label coarseI = dynFaceCells.size();
            dynFaceRestrictAddressing.append(coarseI);
            dynFaceCells.append(localRestrictAddressing[ffi]);
            cellsToCoarseFace.insert(cellPair, coarseI);
        }
    }

    faceCellsHost_.transfer(dynFaceCells);
    faceRestrictAddressingHost_.transfer(dynFaceRestrictAddressing);
    faceCells_ = faceCellsHost_;

    updateAddressing();
}


Foam::cyclicGGAMGInterface::cyclicGGAMGInterface
(
    const label index,
    const gpulduInterfacePtrsList& coarseInterfaces,
    Istream& is
)
:
    GGAMGInterface(index, coarseInterfaces, is),
    neighbPatchID_(readLabel(is)),
    owner_(readBool(is)),
    forwardT_(is),
    reverseT_(is)
{}


// * * * * * * * * * * * * * * * Member Functions  * * * * * * * * * * * * * //

Foam::tmp<Foam::labelgpuField> Foam::cyclicGGAMGInterface::internalFieldTransfer
(
    const Pstream::commsTypes commsType,
    const labelgpuList& iF
) const
{
    const cyclicGGAMGInterface& nbr = neighbPatch();
    const labelgpuList& nbrFaceCells = nbr.gpuFaceCells();

    tmp<labelgpuField> tpnf(new labelgpuField(size()));
    labelgpuField& pnf = tpnf.ref();

    /*forAll(pnf, facei)
    {
        pnf[facei] = iF[nbrFaceCells[facei]];
    }*/

    thrust::copy(thrust::make_permutation_iterator(iF.begin(),nbrFaceCells.begin()),
                 thrust::make_permutation_iterator(iF.begin(),nbrFaceCells.end()),
                 pnf.begin());

    return tpnf;
}


void Foam::cyclicGGAMGInterface::write(Ostream& os) const
{
    GGAMGInterface::write(os);
    os  << token::SPACE << neighbPatchID_
        << token::SPACE << owner_
        << token::SPACE << forwardT_
        << token::SPACE << reverseT_;
}


// ************************************************************************* //
