Commit f5a3cda7 authored by Carter Edwards's avatar Carter Edwards
Browse files

Kokkos: Begin porting core/usecases/hybrid_fem/NonLinear mini-driver to examples/fenl.

Upgrade use case to use containers/, linalg/, example/fixture, and be fully thread parallel.
New thread parallel algorithm to construct sparse matrix graph from element->node connectivity.
parent f8ec13a7
......@@ -171,5 +171,51 @@ create_mirror( const StaticCrsGraph<DataType,Arg1Type,Arg2Type,SizeType > & inpu
//----------------------------------------------------------------------------
//----------------------------------------------------------------------------
namespace Kokkos {
namespace Impl {
template< class GraphType >
struct StaticCrsGraphMaximumEntry {
typedef typename GraphType::device_type device_type ;
typedef typename GraphType::data_type value_type ;
const typename GraphType::entries_type entries ;
StaticCrsGraphMaximumEntry( const GraphType & graph ) : entries( graph.entries ) {}
KOKKOS_INLINE_FUNCTION
void operator()( const unsigned i , value_type & update ) const
{ if ( update < entries(i) ) update = entries(i); }
KOKKOS_INLINE_FUNCTION
void init( value_type & update ) const
{ update = 0 ; }
KOKKOS_INLINE_FUNCTION
void join( volatile value_type & update ,
volatile const value_type & input ) const
{ if ( update < input ) update = input ; }
};
}
template< class DataType, class Arg1Type, class Arg2Type, typename SizeType >
DataType maximum_entry( const StaticCrsGraph< DataType , Arg1Type , Arg2Type , SizeType > & graph )
{
typedef StaticCrsGraph<DataType,Arg1Type,Arg2Type,SizeType> GraphType ;
typedef Impl::StaticCrsGraphMaximumEntry< GraphType > FunctorType ;
DataType result = 0 ;
Kokkos::parallel_reduce( graph.entries.dimension_0(),
FunctorType(graph), result );
return result ;
}
} // namespace Kokkos
//----------------------------------------------------------------------------
//----------------------------------------------------------------------------
#endif /* #ifndef KOKKOS_CRSARRAY_HPP */
......@@ -408,6 +408,11 @@ public:
typename traits::device_type ,
typename traits::memory_traits > const_type ;
typedef View< typename traits::non_const_data_type ,
typename traits::array_layout ,
typename traits::device_type ,
typename traits::memory_traits > non_const_type ;
typedef View< typename traits::non_const_data_type ,
typename traits::array_layout ,
typename traits::device_type::host_mirror_device_type ,
......
......@@ -46,7 +46,7 @@
#if defined( __CUDACC__ )
#include <cusparse.h>
#include <cusparse_v2.h>
#include <Kokkos_Cuda.hpp>
namespace Kokkos {
......
/*
//@HEADER
// ************************************************************************
//
// Kokkos: Manycore Performance-Portable Multidimensional Arrays
// Copyright (2012) Sandia Corporation
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Redistribution and use in source and binary forms, with or without
// modification, are permitted provided that the following conditions are
// met:
//
// 1. Redistributions of source code must retain the above copyright
// notice, this list of conditions and the following disclaimer.
//
// 2. Redistributions in binary form must reproduce the above copyright
// notice, this list of conditions and the following disclaimer in the
// documentation and/or other materials provided with the distribution.
//
// 3. Neither the name of the Corporation nor the names of the
// contributors may be used to endorse or promote products derived from
// this software without specific prior written permission.
//
// THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
//
// Questions? Contact H. Carter Edwards (hcedwar@sandia.gov)
//
// ************************************************************************
//@HEADER
*/
#ifndef KOKKOS_EXAMPLE_WRAP_MPI
#define KOKKOS_EXAMPLE_WRAP_MPI
#include <KokkosCore_config.h>
#include <string>
#if defined( KOKKOS_HAVE_MPI )
#include <mpi.h>
#elif ! defined( KOKKOS_HAVE_MPI )
/* Wrap the the MPI_Comm type and heavily used MPI functions
* to reduce the number of '#if defined( KOKKOS_HAVE_MPI )'
* blocks which have to be sprinkled throughout the examples.
*/
typedef int MPI_Comm ;
inline int MPI_Comm_size( MPI_Comm , int * size ) { *size = 1 ; return 0 ; }
inline int MPI_Comm_rank( MPI_Comm , int * rank ) { *rank = 0 ; return 0 ; }
#endif /* ! defined( KOKKOS_HAVE_MPI ) */
#endif /* #ifndef KOKKOS_EXAMPLE_WRAP_MPI */
INCLUDE(TribitsAddExecutableAndTest)
INCLUDE_DIRECTORIES(${CMAKE_CURRENT_BINARY_DIR})
INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR})
INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR}/../fixture)
INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR}/../common)
SET(SOURCES "")
FILE(GLOB SOURCES *.cpp)
LIST( APPEND SOURCES ../fixture/BoxElemPart.cpp)
IF( Kokkos_ENABLE_Cuda )
# add MPI include dir to nvcc
IF( TPL_ENABLE_MPI )
CUDA_INCLUDE_DIRECTORIES( ${MPI_BASE_DIR}/include )
ENDIF()
CUDA_INCLUDE_DIRECTORIES( ${KokkosCore_INCLUDE_DIRS} )
CUDA_INCLUDE_DIRECTORIES( ${KokkosContainers_INCLUDE_DIRS} )
CUDA_INCLUDE_DIRECTORIES( ${KokkosLinAlg_INCLUDE_DIRS} )
CUDA_COMPILE( OBJECTS_CUDA fenl.cpp.cu )
LIST( APPEND SOURCES ${OBJECTS_CUDA} )
ENDIF()
TRIBITS_ADD_EXECUTABLE(
fenl
SOURCES ${SOURCES}
COMM serial mpi
DEPLIBS kokkoscorelinalg kokkoscontainers kokkoscore
)
#!/bin/bash
#-----------------------------------------------------------------------------
# Simple build script with options
#-----------------------------------------------------------------------------
# Directory for Kokkos
KOKKOS="../../core"
source ${KOKKOS}/src/build_common.sh
# Process command line options and set compilation variables
#
# INC_PATH : required include paths
# CXX : C++ compiler with compiler-specific options
# CXX_SOURCES : C++ files for host
# NVCC : Cuda compiler with compiler-specific options
# NVCC_SOURCES : Cuda sources
#-----------------------------------------------------------------------------
INC_PATH="${INC_PATH} -I../fixture"
CXX_SOURCES="${CXX_SOURCES} ../fixture/BoxElemPart.cpp"
CXX_SOURCES="${CXX_SOURCES} *.cpp"
#-----------------------------------------------------------------------------
# If Cuda compiler set build Cuda source code
if [ -n "${NVCC}" ] ;
then
NVCC_SOURCES="${NVCC_SOURCES} *.cu"
CXX_SOURCES="${CXX_SOURCES}"
echo ${NVCC} ${INC_PATH} ${NVCC_SOURCES}
${NVCC} ${INC_PATH} ${NVCC_SOURCES}
fi
#-----------------------------------------------------------------------------
# Build C++ source code:
EXEC="feint.exe"
echo ${CXX} ${INC_PATH} -o ${EXEC} ${CXX_SOURCES} ${LIB}
rm -f ${EXEC}
${CXX} ${INC_PATH} -o ${EXEC} ${CXX_SOURCES} ${LIB}
rm -f *.o *.a
#-----------------------------------------------------------------------------
/*
// ************************************************************************
//
// Kokkos: Manycore Performance-Portable Multidimensional Arrays
// Copyright (2012) Sandia Corporation
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Redistribution and use in source and binary forms, with or without
// modification, are permitted provided that the following conditions are
// met:
//
// 1. Redistributions of source code must retain the above copyright
// notice, this list of conditions and the following disclaimer.
//
// 2. Redistributions in binary form must reproduce the above copyright
// notice, this list of conditions and the following disclaimer in the
// documentation and/or other materials provided with the distribution.
//
// 3. Neither the name of the Corporation nor the names of the
// contributors may be used to endorse or promote products derived from
// this software without specific prior written permission.
//
// THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
//
// Questions? Contact H. Carter Edwards (hcedwar@sandia.gov)
//
// ************************************************************************
*/
#include <Kokkos_Threads.hpp>
#include <HexElement.hpp>
#include <fenl_impl.hpp>
namespace Kokkos {
namespace Example {
namespace FENL {
template
Perf fenl< Kokkos::Threads , Kokkos::Example::BoxElemPart::ElemLinear >(
MPI_Comm comm ,
const int use_print ,
const int use_trials ,
const int use_atomic ,
const int global_elems[] );
template
Perf fenl< Kokkos::Threads , Kokkos::Example::BoxElemPart::ElemQuadratic >(
MPI_Comm comm ,
const int use_print ,
const int use_trials ,
const int use_atomic ,
const int global_elems[] );
} /* namespace FENL */
} /* namespace Example */
} /* namespace Kokkos */
/*
// ************************************************************************
//
// Kokkos: Manycore Performance-Portable Multidimensional Arrays
// Copyright (2012) Sandia Corporation
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Redistribution and use in source and binary forms, with or without
// modification, are permitted provided that the following conditions are
// met:
//
// 1. Redistributions of source code must retain the above copyright
// notice, this list of conditions and the following disclaimer.
//
// 2. Redistributions in binary form must reproduce the above copyright
// notice, this list of conditions and the following disclaimer in the
// documentation and/or other materials provided with the distribution.
//
// 3. Neither the name of the Corporation nor the names of the
// contributors may be used to endorse or promote products derived from
// this software without specific prior written permission.
//
// THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
//
// Questions? Contact H. Carter Edwards (hcedwar@sandia.gov)
//
// ************************************************************************
*/
#include <Kokkos_Cuda.hpp>
#include <HexElement.hpp>
#include <fenl_impl.hpp>
namespace Kokkos {
namespace Example {
namespace FENL {
template
Perf fenl< Kokkos::Cuda , Kokkos::Example::BoxElemPart::ElemLinear >(
MPI_Comm comm ,
const int use_print ,
const int use_trials ,
const int use_atomic ,
const int global_elems[] );
template
Perf fenl< Kokkos::Cuda , Kokkos::Example::BoxElemPart::ElemQuadratic >(
MPI_Comm comm ,
const int use_print ,
const int use_trials ,
const int use_atomic ,
const int global_elems[] );
} /* namespace FENL */
} /* namespace Example */
} /* namespace Kokkos */
/*
//@HEADER
// ************************************************************************
//
// Kokkos: Manycore Performance-Portable Multidimensional Arrays
// Copyright (2012) Sandia Corporation
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Redistribution and use in source and binary forms, with or without
// modification, are permitted provided that the following conditions are
// met:
//
// 1. Redistributions of source code must retain the above copyright
// notice, this list of conditions and the following disclaimer.
//
// 2. Redistributions in binary form must reproduce the above copyright
// notice, this list of conditions and the following disclaimer in the
// documentation and/or other materials provided with the distribution.
//
// 3. Neither the name of the Corporation nor the names of the
// contributors may be used to endorse or promote products derived from
// this software without specific prior written permission.
//
// THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
//
// Questions? Contact H. Carter Edwards (hcedwar@sandia.gov)
//
// ************************************************************************
//@HEADER
*/
#ifndef KOKKOS_EXAMPLE_FENL_HPP
#define KOKKOS_EXAMPLE_FENL_HPP
#include <stdlib.h>
#include <BoxElemPart.hpp>
#include <WrapMPI.hpp>
namespace Kokkos {
namespace Example {
namespace FENL {
struct Perf {
size_t global_elem_count ;
size_t global_node_count ;
size_t newton_iter_count ;
size_t cg_iter_count ;
double graph_time ;
double elem_time ;
double fill_time ;
double bc_time ;
double cg_time ;
double error_max ;
Perf()
: global_elem_count(0)
, global_node_count(0)
, newton_iter_count(0)
, cg_iter_count(0)
, graph_time(0)
, elem_time(0)
, fill_time(0)
, bc_time(0)
, cg_time(0)
, error_max(0)
{}
};
template < class Device , BoxElemPart::ElemOrder ElemOrder >
Perf fenl(
MPI_Comm comm ,
const int use_print ,
const int use_trials ,
const int use_atomic ,
const int global_elems[] );
} /* namespace FENL */
} /* namespace Example */
} /* namespace Kokkos */
#endif /* #ifndef KOKKOS_EXAMPLE_FENL_HPP */
This diff is collapsed.
/*
//@HEADER
// ************************************************************************
//
// Kokkos: Manycore Performance-Portable Multidimensional Arrays
// Copyright (2012) Sandia Corporation
//
// Under the terms of Contract DE-AC04-94AL85000 with Sandia Corporation,
// the U.S. Government retains certain rights in this software.
//
// Redistribution and use in source and binary forms, with or without
// modification, are permitted provided that the following conditions are
// met:
//
// 1. Redistributions of source code must retain the above copyright
// notice, this list of conditions and the following disclaimer.
//
// 2. Redistributions in binary form must reproduce the above copyright
// notice, this list of conditions and the following disclaimer in the
// documentation and/or other materials provided with the distribution.
//
// 3. Neither the name of the Corporation nor the names of the
// contributors may be used to endorse or promote products derived from
// this software without specific prior written permission.
//
// THIS SOFTWARE IS PROVIDED BY SANDIA CORPORATION "AS IS" AND ANY
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL SANDIA CORPORATION OR THE
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
// LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
// NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
// SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
//
// Questions? Contact H. Carter Edwards (hcedwar@sandia.gov)
//
// ************************************************************************
//@HEADER
*/
#ifndef KOKKOS_EXAMPLE_FENL_IMPL_HPP
#define KOKKOS_EXAMPLE_FENL_IMPL_HPP
#include <math.h>
#include <Kokkos_UnorderedMap.hpp>
#include <Kokkos_StaticCrsGraph.hpp>
#include <Kokkos_CrsMatrix.hpp>
#include <BoxElemFixture.hpp>
#include <fenl.hpp>
#include <fenlFunctors.hpp>
#include <impl/Kokkos_Timer.hpp>
//----------------------------------------------------------------------------
namespace Kokkos {
namespace Example {
namespace FENL {
inline
double maximum( MPI_Comm comm , double local )
{
double global = local ;
#if defined( KOKKOS_HAVE_MPI )
MPI_Allreduce( & local , & global , 1 , MPI_DOUBLE , MPI_MAX , comm );
#endif
return global ;
}
} /* namespace FENL */
} /* namespace Example */
} /* namespace Kokkos */
//----------------------------------------------------------------------------
namespace Kokkos {
namespace Example {
namespace FENL {
class ManufacturedSolution {
public:
// Manufactured solution for one dimensional nonlinear PDE
//
// -K T_zz + T^2 = 0 ; T(zmin) = T_zmin ; T(zmax) = T_zmax
//
// Has an analytic solution of the form:
//
// T(z) = ( a ( z - zmin ) + b )^(-2) where K = 1 / ( 6 a^2 )
//
// Given T_0 and T_L compute K for this analytic solution.
//
// Two analytic solutions:
//
// Solution with singularity:
// , a( ( 1.0 / sqrt(T_zmax) + 1.0 / sqrt(T_zmin) ) / ( zmax - zmin ) )
// , b( -1.0 / sqrt(T_zmin) )
//
// Solution without singularity:
// , a( ( 1.0 / sqrt(T_zmax) - 1.0 / sqrt(T_zmin) ) / ( zmax - zmin ) )
// , b( 1.0 / sqrt(T_zmin) )
const double zmin ;
const double zmax ;
const double T_zmin ;
const double T_zmax ;
const double a ;
const double b ;
const double K ;
ManufacturedSolution( const double arg_zmin ,
const double arg_zmax ,
const double arg_T_zmin ,
const double arg_T_zmax )
: zmin( arg_zmin )
, zmax( arg_zmax )
, T_zmin( arg_T_zmin )
, T_zmax( arg_T_zmax )
, a( ( 1.0 / sqrt(T_zmax) - 1.0 / sqrt(T_zmin) ) / ( zmax - zmin ) )
, b( 1.0 / sqrt(T_zmin) )
, K( 1.0 / ( 6.0 * a * a ) )
{}
double operator()( const double z ) const
{
const double tmp = a * ( z - zmin ) + b ;
return 1.0 / ( tmp * tmp );
}
};
} /* namespace FENL */
} /* namespace Example */