Support for MIC offloading. The process to get TAU working on both th…

…e host and

device at the same time is a little messy--in part because of gaps in Intel's
MIC software stack which should hopefully be resolved in future releases.

See README.mic_offload and examples/mic for the steps to instrument
offload applications.

This does not work with opari. Given the convoluted pomp instrumentation
method this would be a substatial amount of work to extend to MIC offloading.


Former-commit-id: 099cf96afddf8b94ae981c336c59ddc0756260f0

README.mic_offload

@@ -0,0 +1,18 @@
+Configuring TAU for MIC offload applications.
+
+TAU can track both the host and mic sides of a MIC offloaded application. 
+
+1. Configure TAU with both '-arch=mic_linux' and '-arch=x86_64'.
+
+2. Set TAU makefile to x86_64 Makefile. Set TAU option '-optMICOffload', compile
+and link application with the TAU compiler wrappers.
+
+4. Set TAU environment variable: TAU_MIC_OFFLOAD=1, and MIC_PROXY_FS_ROOT=`pwd`
+then run the application.
+
+5. Host profile will be written out as profile.0.0.* and MIC profiles will be
+written out as profile.0.<pid>.*.
+
+NOTE: MIC offloading does not work with opari/pomp. Because it obfuscates the
+instrumentation process the compiler can not tell which portions of the
+measurement library needs to be compiled for the MIC architecture.

examples/mic/Makefile

@@ -0,0 +1,22 @@
+default: offload
+
+sample_native.o: sample.c
+	tau_cc.sh -openmp -mmic -c $< -o $@
+
+sample_host.o: sample.c
+	tau_cc.sh -openmp -no-offload -c $< -o $@
+
+sample.o: sample.c
+	tau_cc.sh -openmp -c $< -o $@
+
+native: sample_native.o
+	tau_cc.sh -openmp -mmic sample_native.o -o $@
+
+host: sample_host.o
+	tau_cc.sh -openmp -no-offload sample_host.o -o $@
+
+offload: sample.o
+	tau_cc.sh -tau_options='-optVerbose -optMICOffload' -openmp sample.o -o $@
+
+clean:
+	rm -f *.o native host offload

examples/mic/sample.c

@@ -0,0 +1,38 @@
+#include <stdio.h>
+#include <math.h>
+
+#define ITERATIONS 1000000000
+
+__attribute__((target(mic))) float compute_pi(int start)
+{
+	 float pi = 0.0f;
+	 float p16 = 1;
+	 int i;
+#pragma omp parallel for private(i) reduction(+:pi)
+	 for (i=start; i<start+ITERATIONS/2; i++)
+	 {	
+	 		p16 = pow(16,i);
+			pi += 1.0/p16 * (4.0/(8*i + 1) - 2.0/(8*i + 4) -
+	          1.0/(8*i + 5) - 1.0/(8*i + 6));
+	 }
+	 return pi;
+}
+
+int main()
+{
+	float pi = 0.0f;
+
+//First offload half the computation to the mic card.
+#pragma offload target (mic)
+	{
+	 	pi = compute_pi(0);
+	}
+
+//Compute the second half on the host.
+	{
+		pi += compute_pi(ITERATIONS/2);
+	}
+
+	printf("Approximation of PI: %.20f.\n", pi);
+
+}

include/Makefile

@@ -1174,10 +1174,8 @@ TAUCOMPILEROPTS=  -optPdtDir="$(PDTDIR)/${PDTARCHDIR}"\
 	-optTauCC="$(TAU_CCOMPILER)" \
 	-optTauIncludes="$(TAU_INCLUDE) $(TAU_MPI_INCLUDE) $(TAU_SHMEM_INCLUDE) $(OPARIINCDIR)" \
 	-optTauDefs="$(TAU_DEFS)" $(TAU_IBM_NOCOMPINST) \
-        -optTauCompile="$(TAU_INCLUDE) $(TAU_MPI_COMPILE_INCLUDE) $(TAU_DEFS) "\
         -optLinking="$(TAU_MPI_FLIBS) $(TAU_LIBS) $(TAU_LDFLAGS) $(TAU_CXXLIBS) $(TAU_SHMEM_LIBS) "\
-        -optSharedLinking="-L$(TAU_LIB_DIR)/shared$(TAU_CONFIG) $(TAU_MPI_FLIBS)  $(TAU_SHMEM_LIBS) $(TAU_EXLIBS) $(TAU_LDFLAGS) $(TAU_CXXLIBS)"\
-	$(TAU_COMPILER_EXTRA_OPTIONS) \
+        -optSharedLinking="-L$(TAU_LIB_DIR)/shared$(TAU_CONFIG) $(TAU_MPI_FLIBS)  $(TAU_SHMEM_LIBS) $(TAU_EXLIBS) $(TAU_LDFLAGS) $(TAU_CXXLIBS) $(TAU_COMPILER_EXTRA_OPTIONS)" \
 				-optMICOffloadLinking="$(MIC_OFFLOAD_LINKING)"\
 				-optMICOffloadSharedLinking="$(MIC_OFFLOAD_SHARED_LINKING)"\
         -optIncludeMemory="$(TAU_INCLUDE_MEMORY)"\

src/Profile/Profiler.cpp

@@ -257,7 +257,6 @@ void Profiler::Start(int tid) {
   RtsLayer::getUSecD(tid, StartTime);	  
   TimeStamp = (x_uint64) StartTime[0]; // USE COUNTER1 for tracing
 
-
   /********************************************************************************/
   /*** Extras ***/
   /********************************************************************************/

src/Profile/TauInit.cpp

@@ -481,6 +481,9 @@ extern "C" int Tau_init_initializeTAU() {
 	Tau_metadata_fillMetaData();
 #endif
 
+  tau_initialized = 1;
+  Tau_global_decr_insideTAU();
+
 #ifdef __MIC__
 	if (TauEnv_get_mic_offload())
 	{
@@ -489,8 +492,6 @@ extern "C" int Tau_init_initializeTAU() {
 	}
 #endif
 
-  tau_initialized = 1;
-  Tau_global_decr_insideTAU();
   return 0;
 }
 

tools/src/tau_compiler.sh

@@ -773,11 +773,6 @@ for arg in "$@" ; do
 			optFixHashIf=$TRUE
 			echoIfDebug "\tFixing Hash-Ifs"
 			;;
-		    -opt*)
-			#Assume any other options should be passed on to the compiler.
-			argsRemaining="$argsRemaining ${arg%% *}"
-			;;
-
 		    -optMICOffloadLinking*)
 			optMICOffloadLinking="${arg#"-optMICOffloadLinking="} $optMICOffloadLinking"
 			echoIfDebug "\tLinking Options are: $optMICOffloadLinking"
@@ -790,6 +785,10 @@ for arg in "$@" ; do
 			optMICOffload=$TRUE
 			echoIfDebug "\tLinking for MIC Offloading"
 			;;
+		    -opt*)
+			#Assume any other options should be passed on to the compiler.
+			argsRemaining="$argsRemaining ${arg%% *}"
+			;;
 
 		esac #end case for parsing script Options
 		;;
@@ -1237,7 +1236,7 @@ if [ $optMICOffload == $TRUE ]; then
 	#	exit 1
 	#fi
 	#hybridLinking="$optLinking -offload-build -offload-ldopts='$optMICLinking'"
-	#echoIfDebug "Hybrid linking options: $hybridLinking"
+	echoIfDebug "MIC offload linking enabled."
 	if [ $optShared == $TRUE ]; then
 		optLinking=$optMICOffloadSharedLinking
 	else