/*---------------------------------------------------------------------------*\
  =========                 |
  \\      /  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) 2020-2021 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 "gpufvPatch.H"
#include "addToRunTimeSelectionTable.H"
#include "gpufvBoundaryMesh.H"
#include "gpufvMesh.H"
#include "primitiveMesh.H"
#include "volgpuFields.H"
#include "surfacegpuFields.H"

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

namespace Foam
{
    defineTypeNameAndDebug(gpufvPatch, 0);
    defineRunTimeSelectionTable(gpufvPatch, polyPatch);
    addToRunTimeSelectionTable(gpufvPatch, gpufvPatch, polyPatch);
}


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

const Foam::fvPatch& Foam::gpufvPatch::lookupPatch(const polyPatch& p)
{
    const fvMesh* meshptr = isA<fvMesh>(p.boundaryMesh().mesh());

    if (!meshptr)
    {
        FatalErrorInFunction
            << "The polyPatch is not attached to a base fvMesh" << nl
            << exit(FatalError);
    }

    return meshptr->boundary()[p.index()];
}


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

Foam::gpufvPatch::gpufvPatch(const polyPatch& p, const gpufvBoundaryMesh& bm)
:
    polyPatch_(p),
    boundaryMesh_(bm)
{}


// * * * * * * * * * * * * * * * * Destructor  * * * * * * * * * * * * * * * //

Foam::gpufvPatch::~gpufvPatch()
{}  // fvBoundaryMesh was forward declared


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

bool Foam::gpufvPatch::constraintType(const word& pt)
{
    return
    (
        fvPatchgpuField<scalar>::patchConstructorTablePtr_
     && fvPatchgpuField<scalar>::patchConstructorTablePtr_->found(pt)
    );
}


Foam::wordList Foam::gpufvPatch::constraintTypes()
{
    const auto& cnstrTable = *polyPatchConstructorTablePtr_;

    wordList cTypes(cnstrTable.size());

    label i = 0;

    forAllConstIters(cnstrTable, iter)
    {
        if (constraintType(iter.key()))
        {
            cTypes[i++] = iter.key();
        }
    }

    cTypes.setSize(i);

    return cTypes;
}


const Foam::labelgpuList& Foam::gpufvPatch::gpuFaceCells() const
{
    return polyPatch_.gpuFaceCells();
}

const Foam::vectorgpuField& Foam::gpufvPatch::Cf() const
{
    return boundaryMesh().mesh().Cf().boundaryField()[index()];
}


Foam::tmp<Foam::vectorgpuField> Foam::gpufvPatch::Cn() const
{
    auto tcc = tmp<vectorgpuField>::New(size());
    auto& cc = tcc.ref();

    const labelgpuList& faceCells = this->gpuFaceCells();

    // get reference to global cell centres
    const vectorgpuField& gcc = boundaryMesh().mesh().hostmesh().gpuCellCentres();

    auto gccIter = thrust::make_permutation_iterator(
        gcc.begin(),
        faceCells.begin()
    );
    thrust::copy(gccIter, gccIter + faceCells.size(), cc.begin());

    return tcc;
}


Foam::tmp<Foam::vectorgpuField> Foam::gpufvPatch::nf() const
{
    return Sf()/magSf();
}


const Foam::vectorgpuField& Foam::gpufvPatch::Sf() const
{
    return boundaryMesh().mesh().Sf().boundaryField()[index()];
}


const Foam::scalargpuField& Foam::gpufvPatch::magSf() const
{
    return boundaryMesh().mesh().magSf().boundaryField()[index()];
}

namespace Foam 
{
struct fvPatchDelta 
{
    template<class Tuple>
    __host__ __device__
    vector operator()(const Tuple& t)
    {
        const vector& S = thrust::get<0>(t);
        const scalar& magS = thrust::get<1>(t);
        const vector& C = thrust::get<2>(t);
        const vector& Cn = thrust::get<3>(t);

        const vector nHat = S/magS;

        return nHat*(nHat & (C - Cn));
    }
};
}

Foam::tmp<Foam::vectorgpuField> Foam::gpufvPatch::delta() const
{
		tmp<vectorgpuField> tout(new vectorgpuField(size()));
		vectorgpuField& out = tout.ref();
	
		const vectorgpuField& S = Sf();
		const scalargpuField& magS = magSf();
		const vectorgpuField& C = Cf();
	
		const vectorgpuField& gcc = boundaryMesh().mesh().hostmesh().gpuCellCentres();
		const labelgpuList& faceCells = this->gpuFaceCells();
	
		auto iter = thrust::make_zip_iterator(thrust::make_tuple(
			S.begin(),
			magS.begin(),
			C.begin(),
			thrust::make_permutation_iterator(
				gcc.begin(),
				faceCells.begin()
			)
		));
		thrust::transform(iter, iter+size(), out.begin(), fvPatchDelta());
		return tout;
}


void Foam::gpufvPatch::makeWeights(scalargpuField& w) const
{
    w = 1.0;
}


void Foam::gpufvPatch::makeDeltaCoeffs(scalargpuField& w) const
{}


void Foam::gpufvPatch::makeNonOrthoDeltaCoeffs(scalargpuField& w) const
{}


void Foam::gpufvPatch::makeNonOrthoCorrVectors(vectorgpuField& w) const
{}


void Foam::gpufvPatch::initMovePoints()
{}


void Foam::gpufvPatch::movePoints()
{}


const Foam::scalargpuField& Foam::gpufvPatch::deltaCoeffs() const
{
    return boundaryMesh().mesh().deltaCoeffs().boundaryField()[index()];
}


const Foam::scalargpuField& Foam::gpufvPatch::weights() const
{
    return boundaryMesh().mesh().weights().boundaryField()[index()];
}


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