[Libreoffice-commits] core.git: 7 commits - sc/Library_scopencl.mk sc/source
Kohei Yoshida
kohei.yoshida at collabora.com
Wed Nov 20 06:40:09 PST 2013
sc/Library_scopencl.mk | 1
sc/source/core/opencl/opencl_device.cxx | 545 +++++++++++++++++++++
sc/source/core/opencl/opencl_device.hxx | 26 +
sc/source/core/opencl/opencl_device_selection.h | 618 ++++++++++++++++++++++++
sc/source/core/opencl/openclwrapper.cxx | 26 -
5 files changed, 1199 insertions(+), 17 deletions(-)
New commits:
commit c24eba06580cbeb1e6a6d0a4d3fe2cfa22abd8a7
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 19:36:24 2013 -0500
C++ style struct / enum declarations without typedef.
Change-Id: Idf9e7ddaf4b28c00222470460478631c37c79a6b
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index 65317c0..bb12593 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -42,11 +42,11 @@ bool bIsInited = false;
bool bIsDeviceSelected = false;
ds_device selectedDevice;
-typedef struct LibreOfficeDeviceScore
+struct LibreOfficeDeviceScore
{
double fTime; // small time means faster device
bool bNoCLErrors; // were there any opencl errors
-} LibreOfficeDeviceScore;
+};
struct LibreOfficeDeviceEvaluationIO
{
@@ -59,14 +59,14 @@ struct LibreOfficeDeviceEvaluationIO
unsigned long outputSize;
};
-typedef struct timer
+struct timer
{
#ifdef _WIN32
LARGE_INTEGER start, stop, frequency;
#else
long long start, stop, frequency;
#endif
-} timer;
+};
const char* source = STRINGIFY(
\n#if defined(KHR_DP_EXTENSION)
diff --git a/sc/source/core/opencl/opencl_device_selection.h b/sc/source/core/opencl/opencl_device_selection.h
index 5761951..b4c60b9 100644
--- a/sc/source/core/opencl/opencl_device_selection.h
+++ b/sc/source/core/opencl/opencl_device_selection.h
@@ -22,7 +22,8 @@
#define DS_DEVICE_NAME_LENGTH 256
-typedef enum {
+enum ds_status
+{
DS_SUCCESS = 0
,DS_INVALID_PROFILE = 1000
,DS_MEMORY_ERROR
@@ -34,28 +35,31 @@ typedef enum {
, DS_PROFILE_FILE_ERROR
, DS_SCORE_SERIALIZER_ERROR
, DS_SCORE_DESERIALIZER_ERROR
-} ds_status;
+};
// device type
-typedef enum {
+enum ds_device_type
+{
DS_DEVICE_NATIVE_CPU = 0
,DS_DEVICE_OPENCL_DEVICE
-} ds_device_type;
+};
-typedef struct {
+struct ds_device
+{
ds_device_type type;
cl_device_id oclDeviceID;
char* oclDeviceName;
char* oclDriverVersion;
void* score; // a pointer to the score data, the content/format is application defined
-} ds_device;
+};
-typedef struct {
+struct ds_profile
+{
unsigned int numDevices;
ds_device* devices;
const char* version;
-} ds_profile;
+};
// deallocate memory used by score
typedef ds_status(* ds_score_release)(void* score);
commit e3ece4b31f734ceffab8fd082c447dac5552d207
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 19:24:23 2013 -0500
Handle a case where the profile initialization fails.
Change-Id: Ifd1ab147926ab4f1d34161840d17fdf1ab3f9810
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index ba0214d..65317c0 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -428,9 +428,16 @@ ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
{
/* Setup */
ds_status status;
- ds_profile* profile;
+ ds_profile* profile = NULL;
status = initDSProfile(&profile, "LibreOffice v0.1");
+ if (!profile)
+ {
+ // failed to initialize profile.
+ selectedDevice.oclDeviceID = NULL;
+ return selectedDevice;
+ }
+
/* Try reading scores from file */
std::string tmpStr(sProfilePath);
const char* fileName = tmpStr.append("sc_opencl_device_profile.dat").c_str();
commit 1283dcb0e20a7f7d81c19818083d813c378c1b4a
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 19:15:40 2013 -0500
Use scoped_ptr here for exception safety.
Normally scoped_ptr is preferred for local heap objects, than direct
new and delete pair.
Change-Id: Ia8783bfe5f9b37a9bd58f08cce89bf85d88199d1
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index 45a28bf..ba0214d 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -18,6 +18,8 @@
#include <iostream>
#include <sstream>
#include <vector>
+#include <boost/scoped_ptr.hpp>
+
#include "opencl_device.hxx"
@@ -446,7 +448,7 @@ ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
if (!bForceSelection) LOG_PRINTF("[DS] Profile file not available (" << fileName << "); performing profiling.");
/* Populate input data for micro-benchmark */
- LibreOfficeDeviceEvaluationIO* testData = new LibreOfficeDeviceEvaluationIO;
+ boost::scoped_ptr<LibreOfficeDeviceEvaluationIO> testData(new LibreOfficeDeviceEvaluationIO);
testData->inputSize = INPUTSIZE;
testData->outputSize = OUTPUTSIZE;
testData->input0.resize(testData->inputSize);
@@ -454,12 +456,12 @@ ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
testData->input2.resize(testData->inputSize);
testData->input3.resize(testData->inputSize);
testData->output.resize(testData->outputSize);
- populateInput(testData);
+ populateInput(testData.get());
/* Perform evaluations */
unsigned int numUpdates;
- status = profileDevices(profile, DS_EVALUATE_ALL, evaluateScoreForDevice, (void*)testData, &numUpdates);
- delete testData;
+ status = profileDevices(profile, DS_EVALUATE_ALL, evaluateScoreForDevice, (void*)testData.get(), &numUpdates);
+
if (DS_SUCCESS == status)
{
/* Write scores to file */
commit a190ccb39ada9c639ae32529bb1b2ba4b9d8cf9d
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 19:11:50 2013 -0500
Safer to use std::vector than heap array objects.
This change also fixes memory leaks.
Change-Id: I68a02a6de0b3d1f125e4c6cd164980a92e0fcbf0
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index 8929221..45a28bf 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -17,6 +17,7 @@
#include <float.h>
#include <iostream>
#include <sstream>
+#include <vector>
#include "opencl_device.hxx"
@@ -45,16 +46,16 @@ typedef struct LibreOfficeDeviceScore
bool bNoCLErrors; // were there any opencl errors
} LibreOfficeDeviceScore;
-typedef struct LibreOfficeDeviceEvaluationIO
+struct LibreOfficeDeviceEvaluationIO
{
- double* input0;
- double* input1;
- double* input2;
- double* input3;
- double* output;
+ std::vector<double> input0;
+ std::vector<double> input1;
+ std::vector<double> input2;
+ std::vector<double> input3;
+ std::vector<double> output;
unsigned long inputSize;
unsigned long outputSize;
-} LibreOfficeDeviceEvaluationIO;
+};
typedef struct timer
{
@@ -164,10 +165,10 @@ double random(double min, double max)
void populateInput(LibreOfficeDeviceEvaluationIO* testData)
{
srand((unsigned int)time(NULL));
- double* input0 = testData->input0;
- double* input1 = testData->input1;
- double* input2 = testData->input2;
- double* input3 = testData->input3;
+ double* input0 = &testData->input0[0];
+ double* input1 = &testData->input1[0];
+ double* input2 = &testData->input2[0];
+ double* input3 = &testData->input3[0];
for (unsigned long i = 0; i < testData->inputSize; i++)
{
input0[i] = random(0, i);
@@ -295,15 +296,15 @@ ds_status evaluateScoreForDevice(ds_device* device, void* evalData)
LibreOfficeDeviceEvaluationIO* testData = (LibreOfficeDeviceEvaluationIO*)evalData;
clKernel = clCreateKernel(clProgram, "DynamicKernel", &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateKernel");
- cl_mem clResult = clCreateBuffer(clContext, CL_MEM_WRITE_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->outputSize, testData->output, &clStatus);
+ cl_mem clResult = clCreateBuffer(clContext, CL_MEM_WRITE_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->outputSize, &testData->output[0], &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clResult");
- cl_mem clInput0 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input0, &clStatus);
+ cl_mem clInput0 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, &testData->input0[0], &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput0");
- cl_mem clInput1 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input1, &clStatus);
+ cl_mem clInput1 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, &testData->input1[0], &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput1");
- cl_mem clInput2 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input2, &clStatus);
+ cl_mem clInput2 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, &testData->input2[0], &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput2");
- cl_mem clInput3 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input3, &clStatus);
+ cl_mem clInput3 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, &testData->input3[0], &clStatus);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput3");
clStatus = clSetKernelArg(clKernel, 0, sizeof(cl_mem), (void*)&clResult);
DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clResult");
@@ -448,21 +449,16 @@ ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
LibreOfficeDeviceEvaluationIO* testData = new LibreOfficeDeviceEvaluationIO;
testData->inputSize = INPUTSIZE;
testData->outputSize = OUTPUTSIZE;
- testData->input0 = new double[testData->inputSize];
- testData->input1 = new double[testData->inputSize];
- testData->input2 = new double[testData->inputSize];
- testData->input3 = new double[testData->inputSize];
- testData->output = new double[testData->outputSize];
+ testData->input0.resize(testData->inputSize);
+ testData->input1.resize(testData->inputSize);
+ testData->input2.resize(testData->inputSize);
+ testData->input3.resize(testData->inputSize);
+ testData->output.resize(testData->outputSize);
populateInput(testData);
/* Perform evaluations */
unsigned int numUpdates;
status = profileDevices(profile, DS_EVALUATE_ALL, evaluateScoreForDevice, (void*)testData, &numUpdates);
- delete testData->output;
- delete testData->input3;
- delete testData->input2;
- delete testData->input1;
- delete testData->input0;
delete testData;
if (DS_SUCCESS == status)
{
commit 44f4b3daa40bbe5d93832e9c545fb0724f965a5e
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 16:59:35 2013 -0500
Remove compiler warnings.
Change-Id: I5f6f50e941031284fe519999a2fddede9753abea
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index 2b103a3..8929221 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -228,7 +228,7 @@ ds_status evaluateScoreForDevice(ds_device* device, void* evalData)
bool bKhrFp64Flag = false;
bool bAmdFp64Flag = false;
- const char* buildOption;
+ const char* buildOption = NULL;
std::string tmpStr("-Dfp_t=double -Dfp_t4=double4 -Dfp_t16=double16 -DINPUTSIZE=");
tmpStr.append(dynamic_cast<std::ostringstream&>(std::ostringstream() << std::dec << INPUTSIZE).str());
@@ -495,7 +495,7 @@ ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
char* overrideDeviceStr = getenv("SC_OPENCL_DEVICE_OVERRIDE");
if (NULL != overrideDeviceStr)
{
- unsigned int overrideDeviceIdx = matchDevice(profile, overrideDeviceStr);
+ int overrideDeviceIdx = matchDevice(profile, overrideDeviceStr);
if (-1 != overrideDeviceIdx)
{
LOG_PRINTF("[DS] Overriding Device Selection (SC_OPENCL_DEVICE_OVERRIDE=" << overrideDeviceStr << ").");
diff --git a/sc/source/core/opencl/opencl_device_selection.h b/sc/source/core/opencl/opencl_device_selection.h
index 6e4eb58..5761951 100644
--- a/sc/source/core/opencl/opencl_device_selection.h
+++ b/sc/source/core/opencl/opencl_device_selection.h
@@ -59,7 +59,7 @@ typedef struct {
// deallocate memory used by score
typedef ds_status(* ds_score_release)(void* score);
-static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr)
+inline ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr)
{
ds_status status = DS_SUCCESS;
if (profile != NULL)
@@ -82,7 +82,7 @@ static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr)
}
-static ds_status initDSProfile(ds_profile** p, const char* version)
+inline ds_status initDSProfile(ds_profile** p, const char* version)
{
int numDevices;
cl_uint numPlatforms;
@@ -197,8 +197,8 @@ typedef enum {
, DS_EVALUATE_NEW_ONLY
} ds_evaluation_type;
-static ds_status profileDevices(ds_profile* profile, const ds_evaluation_type type
- , ds_perf_evaluator evaluator, void* evaluatorData, unsigned int* numUpdates)
+inline ds_status profileDevices(ds_profile* profile, const ds_evaluation_type type,
+ ds_perf_evaluator evaluator, void* evaluatorData, unsigned int* numUpdates)
{
ds_status status = DS_SUCCESS;
unsigned int i;
@@ -259,7 +259,7 @@ static ds_status profileDevices(ds_profile* profile, const ds_evaluation_type ty
typedef ds_status(* ds_score_serializer)(ds_device* device, void** serializedScore, unsigned int* serializedScoreSize);
-static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer, const char* file)
+inline ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer, const char* file)
{
ds_status status = DS_SUCCESS;
FILE* profileFile = NULL;
@@ -338,7 +338,7 @@ static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer ser
}
-static ds_status readProFile(const char* fileName, char** content, size_t* contentSize)
+inline ds_status readProFile(const char* fileName, char** content, size_t* contentSize)
{
FILE* input = NULL;
size_t size = 0;
@@ -370,7 +370,7 @@ static ds_status readProFile(const char* fileName, char** content, size_t* conte
}
-static const char* findString(const char* contentStart, const char* contentEnd, const char* string)
+inline const char* findString(const char* contentStart, const char* contentEnd, const char* string)
{
size_t stringLength;
const char* currentPosition;
@@ -397,7 +397,7 @@ static const char* findString(const char* contentStart, const char* contentEnd,
typedef ds_status(* ds_score_deserializer)(ds_device* device, const unsigned char* serializedScore, unsigned int serializedScoreSize);
-static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer, const char* file)
+inline ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer, const char* file)
{
ds_status status = DS_SUCCESS;
@@ -436,7 +436,7 @@ static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer
}
versionStringLength = strlen(profile->version);
- if (versionStringLength != (dataEnd - dataStart)
+ if (versionStringLength != static_cast<size_t>(dataEnd - dataStart)
|| strncmp(profile->version, dataStart, versionStringLength) != 0)
{
// version mismatch
@@ -538,8 +538,8 @@ static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer
actualDeviceNameLength = strlen(profile->devices[i].oclDeviceName);
driverVersionLength = strlen(profile->devices[i].oclDriverVersion);
- if (actualDeviceNameLength == (deviceNameEnd - deviceNameStart)
- && driverVersionLength == (deviceDriverEnd - deviceDriverStart)
+ if (actualDeviceNameLength == static_cast<size_t>(deviceNameEnd - deviceNameStart)
+ && driverVersionLength == static_cast<size_t>(deviceDriverEnd - deviceDriverStart)
&& strncmp(profile->devices[i].oclDeviceName, deviceNameStart, actualDeviceNameLength) == 0
&& strncmp(profile->devices[i].oclDriverVersion, deviceDriverStart, driverVersionLength) == 0)
{
@@ -594,7 +594,7 @@ cleanup:
return status;
}
-static ds_status getNumDeviceWithEmptyScore(ds_profile* profile, unsigned int* num)
+inline ds_status getNumDeviceWithEmptyScore(ds_profile* profile, unsigned int* num)
{
unsigned int i;
if (profile == NULL || num == NULL) return DS_MEMORY_ERROR;
@@ -603,7 +603,7 @@ static ds_status getNumDeviceWithEmptyScore(ds_profile* profile, unsigned int* n
{
if (profile->devices[i].score == NULL)
{
- *num++;
+ (*num)++;
}
}
return DS_SUCCESS;
commit 5abc2c621f93c11617f7119e893168cb8171afe4
Author: Kohei Yoshida <kohei.yoshida at collabora.com>
Date: Tue Nov 19 16:46:29 2013 -0500
Get it to build on Linux.
Change-Id: Ife109b8a123564c5d8e43c793fc6c9de9b9f2090
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
index a1532ba..2b103a3 100644
--- a/sc/source/core/opencl/opencl_device.cxx
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -7,8 +7,6 @@
* file, You can obtain one at http://mozilla.org/MPL/2.0/.
*/
-#pragma warning (disable : 4996)
-
#ifdef _WIN32
#include <Windows.h>
#else
@@ -16,6 +14,7 @@
#endif
#include <time.h>
#include <math.h>
+#include <float.h>
#include <iostream>
#include <sstream>
#include "opencl_device.hxx"
commit 1989677db7492ea2e91efa9ec6b12578106d6e9f
Author: Jagan Lokanatha <Jagan.Lokanatha at amd.com>
Date: Tue Nov 19 16:36:00 2013 -0500
Enable workload-based device selection in OpenCL.
Change-Id: I8af49ccf498cafb48f9b82fabc4910c754ba0f96
diff --git a/sc/Library_scopencl.mk b/sc/Library_scopencl.mk
index 2c0ac16..f4d400b 100644
--- a/sc/Library_scopencl.mk
+++ b/sc/Library_scopencl.mk
@@ -37,6 +37,7 @@ $(eval $(call gb_Library_use_libraries,scopencl,\
$(eval $(call gb_Library_add_exception_objects,scopencl,\
sc/source/core/opencl/formulagroupcl \
sc/source/core/opencl/openclwrapper \
+ sc/source/core/opencl/opencl_device \
sc/source/core/opencl/opbase \
sc/source/core/opencl/op_financial \
sc/source/core/opencl/op_database \
diff --git a/sc/source/core/opencl/opencl_device.cxx b/sc/source/core/opencl/opencl_device.cxx
new file mode 100644
index 0000000..a1532ba
--- /dev/null
+++ b/sc/source/core/opencl/opencl_device.cxx
@@ -0,0 +1,541 @@
+/* -*- Mode: C++; tab-width: 4; indent-tabs-mode: nil; c-basic-offset: 4 -*- */
+/*
+ * This file is part of the LibreOffice project.
+ *
+ * This Source Code Form is subject to the terms of the Mozilla Public
+ * License, v. 2.0. If a copy of the MPL was not distributed with this
+ * file, You can obtain one at http://mozilla.org/MPL/2.0/.
+ */
+
+#pragma warning (disable : 4996)
+
+#ifdef _WIN32
+#include <Windows.h>
+#else
+#include <sys/time.h>
+#endif
+#include <time.h>
+#include <math.h>
+#include <iostream>
+#include <sstream>
+#include "opencl_device.hxx"
+
+
+#define INPUTSIZE 256*40
+#define OUTPUTSIZE 256*40
+
+#define STRINGIFY(...) #__VA_ARGS__"\n"
+#define LOG_PRINTF(x) (std::cout << x << std::endl)
+//#define LOG_PRINTF(x)
+
+#define DS_CHECK_STATUS(status, name) \
+ if (CL_SUCCESS != status) \
+ { \
+ LOG_PRINTF("[OCL] Error code is " << status << " at " << name); \
+ }
+
+namespace sc { namespace OpenCLDevice {
+
+bool bIsInited = false;
+bool bIsDeviceSelected = false;
+ds_device selectedDevice;
+
+typedef struct LibreOfficeDeviceScore
+{
+ double fTime; // small time means faster device
+ bool bNoCLErrors; // were there any opencl errors
+} LibreOfficeDeviceScore;
+
+typedef struct LibreOfficeDeviceEvaluationIO
+{
+ double* input0;
+ double* input1;
+ double* input2;
+ double* input3;
+ double* output;
+ unsigned long inputSize;
+ unsigned long outputSize;
+} LibreOfficeDeviceEvaluationIO;
+
+typedef struct timer
+{
+#ifdef _WIN32
+ LARGE_INTEGER start, stop, frequency;
+#else
+ long long start, stop, frequency;
+#endif
+} timer;
+
+const char* source = STRINGIFY(
+\n#if defined(KHR_DP_EXTENSION)
+\n#pragma OPENCL EXTENSION cl_khr_fp64 : enable
+\n#elif defined(AMD_DP_EXTENSION)
+\n#pragma OPENCL EXTENSION cl_amd_fp64 : enable
+\n#endif
+ \n
+ int isNan(fp_t a) { return a != a; }
+ fp_t fsum(fp_t a, fp_t b) { return a + b; }
+
+ fp_t fAverage(__global fp_t* input)
+{
+ fp_t sum = 0;
+ int count = 0;
+ for (int i = 0; i < INPUTSIZE; i++)
+ {
+ if (!isNan(input[i]))
+ {
+ sum = fsum(input[i], sum);
+ count += 1;
+ }
+ }
+ return sum / (fp_t)count;
+}
+ fp_t fMin(__global fp_t* input)
+{
+ fp_t min = MAXFLOAT;
+ for (int i = 0; i < INPUTSIZE; i++)
+ {
+ if (!isNan(input[i]))
+ {
+ min = fmin(input[i], min);
+ }
+ }
+ return min;
+}
+ fp_t fSoP(__global fp_t* input0, __global fp_t* input1)
+{
+ fp_t sop = 0.0;
+ for (int i = 0; i < INPUTSIZE; i++)
+ {
+ sop += (isNan(input0[i]) ? 0 : input0[i]) * (isNan(input1[i]) ? 0 : input1[i]);
+ }
+ return sop;
+}
+ __kernel void DynamicKernel(
+ __global fp_t* result, __global fp_t* input0, __global fp_t* input1, __global fp_t* input2, __global fp_t* input3)
+{
+ int gid0 = get_global_id(0);
+ fp_t tmp0 = fAverage(input0);
+ fp_t tmp1 = fMin(input1) * fSoP(input2, input3);
+ result[gid0] = fsum(tmp0, tmp1);
+}
+ );
+
+size_t sourceSize[] = { strlen(source) };
+
+/*************************************************************************/
+/* INTERNAL FUNCTIONS */
+/*************************************************************************/
+/* Timer functions - start timer */
+void timerStart(timer* mytimer)
+{
+#ifdef _WIN32
+ QueryPerformanceCounter(&mytimer->start);
+#else
+ struct timespec s;
+ clock_gettime(CLOCK_MONOTONIC, &s);
+ mytimer->start = (long long)s.tv_sec * (long long)1.0E6 + (long long)s.tv_nsec / (long long)1.0E3;
+#endif
+}
+
+/* Timer functions - stop timer and return difference */
+double timerStop(timer* mytimer)
+{
+#ifdef _WIN32
+ QueryPerformanceCounter(&mytimer->stop);
+ QueryPerformanceFrequency(&mytimer->frequency);
+ double time = ((double)(mytimer->stop.QuadPart - mytimer->start.QuadPart) / mytimer->frequency.QuadPart);
+#else
+ struct timespec s;
+ clock_gettime(CLOCK_MONOTONIC, &s);
+ mytimer->stop = (long long)s.tv_sec * (long long)1.0E6 + (long long)s.tv_nsec / (long long)1.0E3;
+ mytimer->frequency = (long long)1.0E6;
+ double time = ((double)(mytimer->stop - mytimer->start) / mytimer->frequency);
+#endif
+ return time;
+}
+
+/* Random number generator */
+double random(double min, double max)
+{
+ return floor(((double)rand() / ((unsigned int)RAND_MAX + 1)) * (max - min + 1) + min);
+}
+
+/* Populate input */
+void populateInput(LibreOfficeDeviceEvaluationIO* testData)
+{
+ srand((unsigned int)time(NULL));
+ double* input0 = testData->input0;
+ double* input1 = testData->input1;
+ double* input2 = testData->input2;
+ double* input3 = testData->input3;
+ for (unsigned long i = 0; i < testData->inputSize; i++)
+ {
+ input0[i] = random(0, i);
+ input1[i] = random(0, i);
+ input2[i] = random(0, i);
+ input3[i] = random(0, i);
+ }
+}
+/* Encode score object as byte string */
+ds_status serializeScore(ds_device* device, void** serializedScore, unsigned int* serializedScoreSize)
+{
+ *serializedScoreSize = sizeof(LibreOfficeDeviceScore);
+ *serializedScore = (void*)new unsigned char[*serializedScoreSize];
+ memcpy(*serializedScore, device->score, *serializedScoreSize);
+ return DS_SUCCESS;
+}
+
+/* Parses byte string and stores in score object */
+ds_status deserializeScore(ds_device* device, const unsigned char* serializedScore, unsigned int serializedScoreSize)
+{
+ // check that serializedScoreSize == sizeof(LibreOfficeDeviceScore);
+ device->score = new LibreOfficeDeviceScore;
+ memcpy(device->score, serializedScore, serializedScoreSize);
+ return DS_SUCCESS;
+}
+
+/* Releases memory held by score */
+ds_status releaseScore(void* score)
+{
+ if (NULL != score)
+ {
+ delete (LibreOfficeDeviceScore*)score;
+ }
+ return DS_SUCCESS;
+}
+
+/* Evaluate devices */
+ds_status evaluateScoreForDevice(ds_device* device, void* evalData)
+{
+ if (DS_DEVICE_OPENCL_DEVICE == device->type)
+ {
+ /* Evaluating an OpenCL device */
+ LOG_PRINTF("[DS] Device: \"" << device->oclDeviceName << "\" (OpenCL) evaluation...");
+ cl_int clStatus;
+ cl_context clContext;
+ cl_command_queue clQueue;
+ cl_program clProgram;
+ cl_kernel clKernel;
+
+ /* Check for 64-bit float extensions */
+ size_t aDevExtInfoSize = 0;
+ clStatus = clGetDeviceInfo(device->oclDeviceID, CL_DEVICE_EXTENSIONS, 0, NULL, &aDevExtInfoSize);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clGetDeviceInfo");
+
+ char* aExtInfo = new char[aDevExtInfoSize];
+ clStatus = clGetDeviceInfo(device->oclDeviceID, CL_DEVICE_EXTENSIONS, sizeof(char) * aDevExtInfoSize, aExtInfo, NULL);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clGetDeviceInfo");
+
+ bool bKhrFp64Flag = false;
+ bool bAmdFp64Flag = false;
+ const char* buildOption;
+ std::string tmpStr("-Dfp_t=double -Dfp_t4=double4 -Dfp_t16=double16 -DINPUTSIZE=");
+ tmpStr.append(dynamic_cast<std::ostringstream&>(std::ostringstream() << std::dec << INPUTSIZE).str());
+
+ if ((std::string(aExtInfo)).find("cl_khr_fp64") != std::string::npos)
+ {
+ bKhrFp64Flag = true;
+ //buildOption = "-D KHR_DP_EXTENSION -Dfp_t=double -Dfp_t4=double4 -Dfp_t16=double16";
+ tmpStr.append(" -DKHR_DP_EXTENSION");
+ buildOption = tmpStr.c_str();
+ }
+ else if ((std::string(aExtInfo)).find("cl_amd_fp64") != std::string::npos)
+ {
+ bAmdFp64Flag = true;
+ //buildOption = "-D AMD_DP_EXTENSION -Dfp_t=double -Dfp_t4=double4 -Dfp_t16=double16";
+ tmpStr.append(" -DAMD_DP_EXTENSION");
+ buildOption = tmpStr.c_str();
+ }
+ delete[] aExtInfo;
+
+ if (!bKhrFp64Flag && !bAmdFp64Flag)
+ {
+ /* No 64-bit float support */
+ device->score = (void*)new LibreOfficeDeviceScore;
+ ((LibreOfficeDeviceScore*)device->score)->fTime = DBL_MAX;
+ ((LibreOfficeDeviceScore*)device->score)->bNoCLErrors = true;
+ }
+ else
+ {
+ /* 64-bit float support present */
+
+ /* Create context and command queue */
+ clContext = clCreateContext(NULL, 1, &device->oclDeviceID, NULL, NULL, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateContext");
+ clQueue = clCreateCommandQueue(clContext, device->oclDeviceID, 0, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateCommandQueue");
+
+ /* Build program */
+ clProgram = clCreateProgramWithSource(clContext, 1, &source, sourceSize, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateProgramWithSource");
+ clStatus = clBuildProgram(clProgram, 1, &device->oclDeviceID, buildOption, NULL, NULL);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clBuildProgram");
+ if (CL_SUCCESS != clStatus)
+ {
+ /* Build program failed */
+ size_t length;
+ char* buildLog;
+ clStatus = clGetProgramBuildInfo(clProgram, device->oclDeviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &length);
+ buildLog = (char*)malloc(length);
+ clStatus = clGetProgramBuildInfo(clProgram, device->oclDeviceID, CL_PROGRAM_BUILD_LOG, length, buildLog, &length);
+ LOG_PRINTF("[OCL] Build Errors" << std::endl << buildLog);
+ free(buildLog);
+
+ device->score = (void*)new LibreOfficeDeviceScore;
+ ((LibreOfficeDeviceScore*)device->score)->fTime = DBL_MAX;
+ ((LibreOfficeDeviceScore*)device->score)->bNoCLErrors = false;
+ }
+ else
+ {
+ /* Build program succeeded */
+ timer kernelTime;
+ timerStart(&kernelTime);
+
+ /* Run kernel */
+ LibreOfficeDeviceEvaluationIO* testData = (LibreOfficeDeviceEvaluationIO*)evalData;
+ clKernel = clCreateKernel(clProgram, "DynamicKernel", &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateKernel");
+ cl_mem clResult = clCreateBuffer(clContext, CL_MEM_WRITE_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->outputSize, testData->output, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clResult");
+ cl_mem clInput0 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input0, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput0");
+ cl_mem clInput1 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input1, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput1");
+ cl_mem clInput2 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input2, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput2");
+ cl_mem clInput3 = clCreateBuffer(clContext, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR, sizeof(cl_double) * testData->inputSize, testData->input3, &clStatus);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clCreateBuffer::clInput3");
+ clStatus = clSetKernelArg(clKernel, 0, sizeof(cl_mem), (void*)&clResult);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clResult");
+ clStatus = clSetKernelArg(clKernel, 1, sizeof(cl_mem), (void*)&clInput0);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clInput0");
+ clStatus = clSetKernelArg(clKernel, 2, sizeof(cl_mem), (void*)&clInput1);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clInput1");
+ clStatus = clSetKernelArg(clKernel, 3, sizeof(cl_mem), (void*)&clInput2);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clInput2");
+ clStatus = clSetKernelArg(clKernel, 4, sizeof(cl_mem), (void*)&clInput3);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clSetKernelArg::clInput3");
+ size_t globalWS[1] = { testData->outputSize };
+ size_t localSize[1] = { 64 };
+ clStatus = clEnqueueNDRangeKernel(clQueue, clKernel, 1, 0, globalWS, localSize, 0, NULL, NULL);
+ DS_CHECK_STATUS(clStatus, "evaluateScoreForDevice::clEnqueueNDRangeKernel");
+ clFinish(clQueue);
+ clReleaseMemObject(clInput3);
+ clReleaseMemObject(clInput2);
+ clReleaseMemObject(clInput1);
+ clReleaseMemObject(clInput0);
+ clReleaseMemObject(clResult);
+ clReleaseKernel(clKernel);
+
+ device->score = (void*)new LibreOfficeDeviceScore;
+ ((LibreOfficeDeviceScore*)device->score)->fTime = timerStop(&kernelTime);
+ ((LibreOfficeDeviceScore*)device->score)->bNoCLErrors = true;
+ }
+
+ clReleaseProgram(clProgram);
+ clReleaseCommandQueue(clQueue);
+ clReleaseContext(clContext);
+ }
+ }
+ else
+ {
+ /* Evaluating an Native CPU device */
+ LOG_PRINTF("[DS] Device: \"CPU\" (Native) evaluation...");
+ timer kernelTime;
+ timerStart(&kernelTime);
+
+ LibreOfficeDeviceEvaluationIO* testData = (LibreOfficeDeviceEvaluationIO*)evalData;
+ for (unsigned long j = 0; j < testData->outputSize; j++)
+ {
+ double fAverage = 0.0f;
+ double fMin = DBL_MAX;
+ double fSoP = 0.0f;
+ for (unsigned long i = 0; i < testData->inputSize; i++)
+ {
+ fAverage += testData->input0[i];
+ fMin = ((fMin < testData->input1[i]) ? fMin : testData->input1[i]);
+ fSoP += testData->input2[i] * testData->input3[i];
+ }
+ fAverage /= testData->inputSize;
+ testData->output[j] = fAverage + (fMin * fSoP);
+ }
+
+ device->score = (void*)new LibreOfficeDeviceScore;
+ ((LibreOfficeDeviceScore*)device->score)->fTime = timerStop(&kernelTime);
+ ((LibreOfficeDeviceScore*)device->score)->bNoCLErrors = true;
+ }
+ return DS_SUCCESS;
+}
+
+/* Pick best device */
+ds_status pickBestDevice(ds_profile* profile, int* bestDeviceIdx)
+{
+ double bestScore = DBL_MAX;
+ *bestDeviceIdx = -1;
+
+ for (unsigned int d = 0; d < profile->numDevices; d++)
+ {
+ ds_device device = profile->devices[d];
+ LibreOfficeDeviceScore score = *(LibreOfficeDeviceScore*)device.score;
+ if (DS_DEVICE_OPENCL_DEVICE == device.type)
+ {
+ LOG_PRINTF("[DS] Device[" << d << "] " << device.oclDeviceName << " (OpenCL) score is " << score.fTime);
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Device[" << d << "] CPU (Native) score is " << score.fTime);
+ }
+ if (score.fTime < bestScore)
+ {
+ bestScore = score.fTime;
+ *bestDeviceIdx = d;
+ }
+ }
+ if (DS_DEVICE_OPENCL_DEVICE == profile->devices[*bestDeviceIdx].type)
+ {
+ LOG_PRINTF("[DS] Selected Device[" << *bestDeviceIdx << "]: " << profile->devices[*bestDeviceIdx].oclDeviceName << "(OpenCL).");
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Selected Device[" << *bestDeviceIdx << "]: CPU (Native).");
+ }
+
+ return DS_SUCCESS;
+}
+
+/* Return device ID for matching device name */
+int matchDevice(ds_profile* profile, char* deviceName)
+{
+ int deviceMatch = -1;
+ for (unsigned int d = 0; d < profile->numDevices - 1; d++)
+ {
+ if ((std::string(profile->devices[d].oclDeviceName)).find(deviceName) != std::string::npos) deviceMatch = d;
+ }
+ if (std::string("NATIVE_CPU").find(deviceName) != std::string::npos) deviceMatch = profile->numDevices - 1;
+ return deviceMatch;
+}
+
+/*************************************************************************/
+/* EXTERNAL FUNCTIONS */
+/*************************************************************************/
+ds_device getDeviceSelection(const char* sProfilePath, bool bForceSelection)
+{
+ /* Run only if device is not yet selected */
+ if (!bIsDeviceSelected || bForceSelection)
+ {
+ /* Setup */
+ ds_status status;
+ ds_profile* profile;
+ status = initDSProfile(&profile, "LibreOffice v0.1");
+
+ /* Try reading scores from file */
+ std::string tmpStr(sProfilePath);
+ const char* fileName = tmpStr.append("sc_opencl_device_profile.dat").c_str();
+ if (!bForceSelection)
+ {
+ status = readProfileFromFile(profile, deserializeScore, fileName);
+ }
+ else
+ {
+ status = DS_INVALID_PROFILE;
+ LOG_PRINTF("[DS] Performing forced profiling.");
+ }
+ if (DS_SUCCESS != status)
+ {
+ if (!bForceSelection) LOG_PRINTF("[DS] Profile file not available (" << fileName << "); performing profiling.");
+
+ /* Populate input data for micro-benchmark */
+ LibreOfficeDeviceEvaluationIO* testData = new LibreOfficeDeviceEvaluationIO;
+ testData->inputSize = INPUTSIZE;
+ testData->outputSize = OUTPUTSIZE;
+ testData->input0 = new double[testData->inputSize];
+ testData->input1 = new double[testData->inputSize];
+ testData->input2 = new double[testData->inputSize];
+ testData->input3 = new double[testData->inputSize];
+ testData->output = new double[testData->outputSize];
+ populateInput(testData);
+
+ /* Perform evaluations */
+ unsigned int numUpdates;
+ status = profileDevices(profile, DS_EVALUATE_ALL, evaluateScoreForDevice, (void*)testData, &numUpdates);
+ delete testData->output;
+ delete testData->input3;
+ delete testData->input2;
+ delete testData->input1;
+ delete testData->input0;
+ delete testData;
+ if (DS_SUCCESS == status)
+ {
+ /* Write scores to file */
+ status = writeProfileToFile(profile, serializeScore, fileName);
+ if (DS_SUCCESS == status)
+ {
+ LOG_PRINTF("[DS] Scores written to file (" << fileName << ").");
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Error saving scores to file (" << fileName << "); scores not written to file.");
+ }
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Unable to evaluate performance; scores not written to file.");
+ }
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Profile read from file (" << fileName << ").");
+ }
+
+ /* Pick best device */
+ int bestDeviceIdx;
+ pickBestDevice(profile, &bestDeviceIdx);
+
+ /* Overide if necessary */
+ char* overrideDeviceStr = getenv("SC_OPENCL_DEVICE_OVERRIDE");
+ if (NULL != overrideDeviceStr)
+ {
+ unsigned int overrideDeviceIdx = matchDevice(profile, overrideDeviceStr);
+ if (-1 != overrideDeviceIdx)
+ {
+ LOG_PRINTF("[DS] Overriding Device Selection (SC_OPENCL_DEVICE_OVERRIDE=" << overrideDeviceStr << ").");
+ bestDeviceIdx = overrideDeviceIdx;
+ if (DS_DEVICE_OPENCL_DEVICE == profile->devices[bestDeviceIdx].type)
+ {
+ LOG_PRINTF("[DS] Selected Device[" << bestDeviceIdx << "]: " << profile->devices[bestDeviceIdx].oclDeviceName << " (OpenCL).");
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Selected Device[" << bestDeviceIdx << "]: CPU (Native).");
+ }
+ }
+ else
+ {
+ LOG_PRINTF("[DS] Ignoring invalid SC_OPENCL_DEVICE_OVERRIDE=" << overrideDeviceStr << ").");
+ }
+ }
+
+ /* Final device selection */
+ selectedDevice = profile->devices[bestDeviceIdx];
+ bIsDeviceSelected = true;
+
+ /* Release profile */
+ status = releaseDSProfile(profile, releaseScore);
+ }
+ return selectedDevice;
+}
+
+bool selectedDeviceIsOpenCL(ds_device device)
+{
+ return (DS_DEVICE_OPENCL_DEVICE == device.type);
+}
+
+bool selectedDeviceIsNativeCPU(ds_device device)
+{
+ return (DS_DEVICE_NATIVE_CPU == device.type);
+}
+
+}}
+
+/* vim:set shiftwidth=4 softtabstop=4 expandtab: */
diff --git a/sc/source/core/opencl/opencl_device.hxx b/sc/source/core/opencl/opencl_device.hxx
new file mode 100644
index 0000000..e013943
--- /dev/null
+++ b/sc/source/core/opencl/opencl_device.hxx
@@ -0,0 +1,26 @@
+/* -*- Mode: C++; tab-width: 4; indent-tabs-mode: nil; c-basic-offset: 4 -*- */
+/*
+ * This file is part of the LibreOffice project.
+ *
+ * This Source Code Form is subject to the terms of the Mozilla Public
+ * License, v. 2.0. If a copy of the MPL was not distributed with this
+ * file, You can obtain one at http://mozilla.org/MPL/2.0/.
+ */
+
+#ifndef SC_OPENCL_DEVICE_HXX
+#define SC_OPENCL_DEVICE_HXX
+
+#pragma once
+#include "opencl_device_selection.h"
+
+namespace sc { namespace OpenCLDevice {
+
+ds_device getDeviceSelection(const char* pFileName, bool bForceSelection = false);
+bool selectedDeviceIsOpenCL(ds_device device);
+bool selectedDeviceIsNativeCPU(ds_device device);
+
+}}
+
+#endif
+
+/* vim:set shiftwidth=4 softtabstop=4 expandtab: */
diff --git a/sc/source/core/opencl/opencl_device_selection.h b/sc/source/core/opencl/opencl_device_selection.h
new file mode 100644
index 0000000..6e4eb58
--- /dev/null
+++ b/sc/source/core/opencl/opencl_device_selection.h
@@ -0,0 +1,614 @@
+/* -*- Mode: C++; tab-width: 4; indent-tabs-mode: nil; c-basic-offset: 4 -*- */
+/*
+ * This file is part of the LibreOffice project.
+ *
+ * This Source Code Form is subject to the terms of the Mozilla Public
+ * License, v. 2.0. If a copy of the MPL was not distributed with this
+ * file, You can obtain one at http://mozilla.org/MPL/2.0/.
+ */
+
+#ifndef SC_DEVICE_SELECTION_H
+#define SC_DEVICE_SELECTION_H
+
+
+#ifdef _MSC_VER
+#define _CRT_SECURE_NO_WARNINGS
+#endif
+
+#include <stdlib.h>
+#include <stdio.h>
+#include <string.h>
+#include "clcc/clew.h"
+
+#define DS_DEVICE_NAME_LENGTH 256
+
+typedef enum {
+ DS_SUCCESS = 0
+ ,DS_INVALID_PROFILE = 1000
+ ,DS_MEMORY_ERROR
+ , DS_INVALID_PERF_EVALUATOR_TYPE
+ , DS_INVALID_PERF_EVALUATOR
+ , DS_PERF_EVALUATOR_ERROR
+ , DS_FILE_ERROR
+ , DS_UNKNOWN_DEVICE_TYPE
+ , DS_PROFILE_FILE_ERROR
+ , DS_SCORE_SERIALIZER_ERROR
+ , DS_SCORE_DESERIALIZER_ERROR
+} ds_status;
+
+// device type
+typedef enum {
+ DS_DEVICE_NATIVE_CPU = 0
+ ,DS_DEVICE_OPENCL_DEVICE
+} ds_device_type;
+
+
+typedef struct {
+ ds_device_type type;
+ cl_device_id oclDeviceID;
+ char* oclDeviceName;
+ char* oclDriverVersion;
+ void* score; // a pointer to the score data, the content/format is application defined
+} ds_device;
+
+typedef struct {
+ unsigned int numDevices;
+ ds_device* devices;
+ const char* version;
+} ds_profile;
+
+// deallocate memory used by score
+typedef ds_status(* ds_score_release)(void* score);
+static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr)
+{
+ ds_status status = DS_SUCCESS;
+ if (profile != NULL)
+ {
+ if (profile->devices != NULL && sr != NULL)
+ {
+ unsigned int i;
+ for (i = 0; i < profile->numDevices; i++)
+ {
+ free(profile->devices[i].oclDeviceName);
+ free(profile->devices[i].oclDriverVersion);
+ status = sr(profile->devices[i].score);
+ if (status != DS_SUCCESS) break;
+ }
+ free(profile->devices);
+ }
+ free(profile);
+ }
+ return status;
+}
+
+
+static ds_status initDSProfile(ds_profile** p, const char* version)
+{
+ int numDevices;
+ cl_uint numPlatforms;
+ cl_platform_id* platforms = NULL;
+ cl_device_id* devices = NULL;
+ ds_status status = DS_SUCCESS;
+ ds_profile* profile = NULL;
+ unsigned int next;
+ unsigned int i;
+
+ if (p == NULL) return DS_INVALID_PROFILE;
+
+ profile = (ds_profile*)malloc(sizeof(ds_profile));
+ if (profile == NULL) return DS_MEMORY_ERROR;
+
+ memset(profile, 0, sizeof(ds_profile));
+
+ clGetPlatformIDs(0, NULL, &numPlatforms);
+ if (numPlatforms != 0)
+ {
+ platforms = (cl_platform_id*)malloc(numPlatforms * sizeof(cl_platform_id));
+ if (platforms == NULL)
+ {
+ status = DS_MEMORY_ERROR;
+ goto cleanup;
+ }
+ clGetPlatformIDs(numPlatforms, platforms, NULL);
+ }
+
+ numDevices = 0;
+ for (i = 0; i < (unsigned int)numPlatforms; i++)
+ {
+ cl_uint num;
+ clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_ALL, 0, NULL, &num);
+ numDevices += num;
+ }
+ if (numDevices != 0)
+ {
+ devices = (cl_device_id*)malloc(numDevices * sizeof(cl_device_id));
+ if (devices == NULL)
+ {
+ status = DS_MEMORY_ERROR;
+ goto cleanup;
+ }
+ }
+
+ profile->numDevices = numDevices + 1; // +1 to numDevices to include the native CPU
+ profile->devices = (ds_device*)malloc(profile->numDevices * sizeof(ds_device));
+ if (profile->devices == NULL)
+ {
+ profile->numDevices = 0;
+ status = DS_MEMORY_ERROR;
+ goto cleanup;
+ }
+ memset(profile->devices, 0, profile->numDevices * sizeof(ds_device));
+
+ next = 0;
+ for (i = 0; i < (unsigned int)numPlatforms; i++)
+ {
+ cl_uint num;
+ unsigned j;
+ clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_ALL, numDevices, devices, &num);
+ for (j = 0; j < num; j++, next++)
+ {
+ char buffer[DS_DEVICE_NAME_LENGTH];
+ size_t length;
+
+ profile->devices[next].type = DS_DEVICE_OPENCL_DEVICE;
+ profile->devices[next].oclDeviceID = devices[j];
+
+ clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
+ , DS_DEVICE_NAME_LENGTH, &buffer, NULL);
+ length = strlen(buffer);
+ profile->devices[next].oclDeviceName = (char*)malloc(length + 1);
+ memcpy(profile->devices[next].oclDeviceName, buffer, length + 1);
+
+ clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
+ , DS_DEVICE_NAME_LENGTH, &buffer, NULL);
+ length = strlen(buffer);
+ profile->devices[next].oclDriverVersion = (char*)malloc(length + 1);
+ memcpy(profile->devices[next].oclDriverVersion, buffer, length + 1);
+ }
+ }
+ profile->devices[next].type = DS_DEVICE_NATIVE_CPU;
+ profile->version = version;
+
+cleanup:
+ if (platforms) free(platforms);
+ if (devices) free(devices);
+ if (status == DS_SUCCESS)
+ {
+ *p = profile;
+ }
+ else
+ {
+ if (profile)
+ {
+ if (profile->devices) free(profile->devices);
+ free(profile);
+ }
+ }
+ return status;
+}
+
+// Pointer to a function that calculates the score of a device (ex: device->score)
+// update the data size of score. The encoding and the format of the score data
+// is implementation defined. The function should return DS_SUCCESS if there's no error to be reported.
+typedef ds_status(* ds_perf_evaluator)(ds_device* device, void* data);
+
+typedef enum {
+ DS_EVALUATE_ALL
+ , DS_EVALUATE_NEW_ONLY
+} ds_evaluation_type;
+
+static ds_status profileDevices(ds_profile* profile, const ds_evaluation_type type
+ , ds_perf_evaluator evaluator, void* evaluatorData, unsigned int* numUpdates)
+{
+ ds_status status = DS_SUCCESS;
+ unsigned int i;
+ unsigned int updates = 0;
+
+ if (profile == NULL)
+ {
+ return DS_INVALID_PROFILE;
+ }
+ if (evaluator == NULL)
+ {
+ return DS_INVALID_PERF_EVALUATOR;
+ }
+
+ for (i = 0; i < profile->numDevices; i++)
+ {
+ ds_status evaluatorStatus;
+
+ switch (type)
+ {
+ case DS_EVALUATE_NEW_ONLY:
+ if (profile->devices[i].score != NULL) break;
+ // else fall through
+ case DS_EVALUATE_ALL:
+ evaluatorStatus = evaluator(profile->devices + i, evaluatorData);
+ if (evaluatorStatus != DS_SUCCESS)
+ {
+ status = evaluatorStatus;
+ return status;
+ }
+ updates++;
+ break;
+ default:
+ return DS_INVALID_PERF_EVALUATOR_TYPE;
+ break;
+ };
+ }
+ if (numUpdates) *numUpdates = updates;
+ return status;
+}
+
+
+#define DS_TAG_VERSION "<version>"
+#define DS_TAG_VERSION_END "</version>"
+#define DS_TAG_DEVICE "<device>"
+#define DS_TAG_DEVICE_END "</device>"
+#define DS_TAG_SCORE "<score>"
+#define DS_TAG_SCORE_END "</score>"
+#define DS_TAG_DEVICE_TYPE "<type>"
+#define DS_TAG_DEVICE_TYPE_END "</type>"
+#define DS_TAG_DEVICE_NAME "<name>"
+#define DS_TAG_DEVICE_NAME_END "</name>"
+#define DS_TAG_DEVICE_DRIVER_VERSION "<driver>"
+#define DS_TAG_DEVICE_DRIVER_VERSION_END "</driver>"
+
+#define DS_DEVICE_NATIVE_CPU_STRING "native_cpu"
+
+
+
+typedef ds_status(* ds_score_serializer)(ds_device* device, void** serializedScore, unsigned int* serializedScoreSize);
+static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer, const char* file)
+{
+ ds_status status = DS_SUCCESS;
+ FILE* profileFile = NULL;
+
+
+ if (profile == NULL) return DS_INVALID_PROFILE;
+
+ profileFile = fopen(file, "wb");
+ if (profileFile == NULL)
+ {
+ status = DS_FILE_ERROR;
+ }
+ else
+ {
+ unsigned int i;
+
+ // write version string
+ fwrite(DS_TAG_VERSION, sizeof(char), strlen(DS_TAG_VERSION), profileFile);
+ fwrite(profile->version, sizeof(char), strlen(profile->version), profileFile);
+ fwrite(DS_TAG_VERSION_END, sizeof(char), strlen(DS_TAG_VERSION_END), profileFile);
+ fwrite("\n", sizeof(char), 1, profileFile);
+
+ for (i = 0; i < profile->numDevices && status == DS_SUCCESS; i++)
+ {
+ void* serializedScore;
+ unsigned int serializedScoreSize;
+
+ fwrite(DS_TAG_DEVICE, sizeof(char), strlen(DS_TAG_DEVICE), profileFile);
+
+ fwrite(DS_TAG_DEVICE_TYPE, sizeof(char), strlen(DS_TAG_DEVICE_TYPE), profileFile);
+ fwrite(&profile->devices[i].type, sizeof(ds_device_type), 1, profileFile);
+ fwrite(DS_TAG_DEVICE_TYPE_END, sizeof(char), strlen(DS_TAG_DEVICE_TYPE_END), profileFile);
+
+ switch (profile->devices[i].type)
+ {
+ case DS_DEVICE_NATIVE_CPU:
+ {
+ // There's no need to emit a device name for the native CPU device.
+ /*
+ fwrite(DS_TAG_DEVICE_NAME, sizeof(char), strlen(DS_TAG_DEVICE_NAME), profileFile);
+ fwrite(DS_DEVICE_NATIVE_CPU_STRING,sizeof(char),strlen(DS_DEVICE_NATIVE_CPU_STRING), profileFile);
+ fwrite(DS_TAG_DEVICE_NAME_END, sizeof(char), strlen(DS_TAG_DEVICE_NAME_END), profileFile);
+ */
+ }
+ break;
+ case DS_DEVICE_OPENCL_DEVICE:
+ {
+ fwrite(DS_TAG_DEVICE_NAME, sizeof(char), strlen(DS_TAG_DEVICE_NAME), profileFile);
+ fwrite(profile->devices[i].oclDeviceName, sizeof(char), strlen(profile->devices[i].oclDeviceName), profileFile);
+ fwrite(DS_TAG_DEVICE_NAME_END, sizeof(char), strlen(DS_TAG_DEVICE_NAME_END), profileFile);
+
+ fwrite(DS_TAG_DEVICE_DRIVER_VERSION, sizeof(char), strlen(DS_TAG_DEVICE_DRIVER_VERSION), profileFile);
+ fwrite(profile->devices[i].oclDriverVersion, sizeof(char), strlen(profile->devices[i].oclDriverVersion), profileFile);
+ fwrite(DS_TAG_DEVICE_DRIVER_VERSION_END, sizeof(char), strlen(DS_TAG_DEVICE_DRIVER_VERSION_END), profileFile);
+ }
+ break;
+ default:
+ status = DS_UNKNOWN_DEVICE_TYPE;
+ break;
+ };
+
+ fwrite(DS_TAG_SCORE, sizeof(char), strlen(DS_TAG_SCORE), profileFile);
+ status = serializer(profile->devices + i, &serializedScore, &serializedScoreSize);
+ if (status == DS_SUCCESS && serializedScore != NULL && serializedScoreSize > 0)
+ {
+ fwrite(serializedScore, sizeof(char), serializedScoreSize, profileFile);
+ free(serializedScore);
+ }
+ fwrite(DS_TAG_SCORE_END, sizeof(char), strlen(DS_TAG_SCORE_END), profileFile);
+ fwrite(DS_TAG_DEVICE_END, sizeof(char), strlen(DS_TAG_DEVICE_END), profileFile);
+ fwrite("\n", sizeof(char), 1, profileFile);
+ }
+ fclose(profileFile);
+ }
+ return status;
+}
+
+
+static ds_status readProFile(const char* fileName, char** content, size_t* contentSize)
+{
+ FILE* input = NULL;
+ size_t size = 0;
+ char* binary = NULL;
+
+ *contentSize = 0;
+ *content = NULL;
+
+ input = fopen(fileName, "rb");
+ if (input == NULL)
+ {
+ return DS_FILE_ERROR;
+ }
+
+ fseek(input, 0L, SEEK_END);
+ size = ftell(input);
+ rewind(input);
+ binary = (char*)malloc(size);
+ if (binary == NULL)
+ {
+ return DS_FILE_ERROR;
+ }
+ fread(binary, sizeof(char), size, input);
+ fclose(input);
+
+ *contentSize = size;
+ *content = binary;
+ return DS_SUCCESS;
+}
+
+
+static const char* findString(const char* contentStart, const char* contentEnd, const char* string)
+{
+ size_t stringLength;
+ const char* currentPosition;
+ const char* found;
+ found = NULL;
+ stringLength = strlen(string);
+ currentPosition = contentStart;
+ for (currentPosition = contentStart; currentPosition < contentEnd; currentPosition++)
+ {
+ if (*currentPosition == string[0])
+ {
+ if (currentPosition + stringLength < contentEnd)
+ {
+ if (strncmp(currentPosition, string, stringLength) == 0)
+ {
+ found = currentPosition;
+ break;
+ }
+ }
+ }
+ }
+ return found;
+}
+
+
+typedef ds_status(* ds_score_deserializer)(ds_device* device, const unsigned char* serializedScore, unsigned int serializedScoreSize);
+static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer, const char* file)
+{
+
+ ds_status status = DS_SUCCESS;
+ char* contentStart = NULL;
+ const char* contentEnd = NULL;
+ size_t contentSize;
+
+ if (profile == NULL) return DS_INVALID_PROFILE;
+
+ status = readProFile(file, &contentStart, &contentSize);
+ if (status == DS_SUCCESS)
+ {
+ const char* currentPosition;
+ const char* dataStart;
+ const char* dataEnd;
+ size_t versionStringLength;
+
+ contentEnd = contentStart + contentSize;
+ currentPosition = contentStart;
+
+
+ // parse the version string
+ dataStart = findString(currentPosition, contentEnd, DS_TAG_VERSION);
+ if (dataStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ dataStart += strlen(DS_TAG_VERSION);
+
+ dataEnd = findString(dataStart, contentEnd, DS_TAG_VERSION_END);
+ if (dataEnd == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+
+ versionStringLength = strlen(profile->version);
+ if (versionStringLength != (dataEnd - dataStart)
+ || strncmp(profile->version, dataStart, versionStringLength) != 0)
+ {
+ // version mismatch
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ currentPosition = dataEnd + strlen(DS_TAG_VERSION_END);
+
+ // parse the device information
+ while (1)
+ {
+ unsigned int i;
+
+ const char* deviceTypeStart;
+ const char* deviceTypeEnd;
+ ds_device_type deviceType;
+
+ const char* deviceNameStart;
+ const char* deviceNameEnd;
+
+ const char* deviceScoreStart;
+ const char* deviceScoreEnd;
+
+ const char* deviceDriverStart;
+ const char* deviceDriverEnd;
+
+ dataStart = findString(currentPosition, contentEnd, DS_TAG_DEVICE);
+ if (dataStart == NULL)
+ {
+ // nothing useful remain, quit...
+ break;
+ }
+ dataStart += strlen(DS_TAG_DEVICE);
+ dataEnd = findString(dataStart, contentEnd, DS_TAG_DEVICE_END);
+ if (dataEnd == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+
+ // parse the device type
+ deviceTypeStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_TYPE);
+ if (deviceTypeStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ deviceTypeStart += strlen(DS_TAG_DEVICE_TYPE);
+ deviceTypeEnd = findString(deviceTypeStart, contentEnd, DS_TAG_DEVICE_TYPE_END);
+ if (deviceTypeEnd == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ memcpy(&deviceType, deviceTypeStart, sizeof(ds_device_type));
+
+
+ // parse the device name
+ if (deviceType == DS_DEVICE_OPENCL_DEVICE)
+ {
+
+ deviceNameStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_NAME);
+ if (deviceNameStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ deviceNameStart += strlen(DS_TAG_DEVICE_NAME);
+ deviceNameEnd = findString(deviceNameStart, contentEnd, DS_TAG_DEVICE_NAME_END);
+ if (deviceNameEnd == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+
+
+ deviceDriverStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION);
+ if (deviceDriverStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ deviceDriverStart += strlen(DS_TAG_DEVICE_DRIVER_VERSION);
+ deviceDriverEnd = findString(deviceDriverStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION_END);
+ if (deviceDriverEnd == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+
+
+ // check if this device is on the system
+ for (i = 0; i < profile->numDevices; i++)
+ {
+ if (profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE)
+ {
+ size_t actualDeviceNameLength;
+ size_t driverVersionLength;
+
+ actualDeviceNameLength = strlen(profile->devices[i].oclDeviceName);
+ driverVersionLength = strlen(profile->devices[i].oclDriverVersion);
+ if (actualDeviceNameLength == (deviceNameEnd - deviceNameStart)
+ && driverVersionLength == (deviceDriverEnd - deviceDriverStart)
+ && strncmp(profile->devices[i].oclDeviceName, deviceNameStart, actualDeviceNameLength) == 0
+ && strncmp(profile->devices[i].oclDriverVersion, deviceDriverStart, driverVersionLength) == 0)
+ {
+
+ deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
+ if (deviceNameStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ deviceScoreStart += strlen(DS_TAG_SCORE);
+ deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
+ status = deserializer(profile->devices + i, (const unsigned char*)deviceScoreStart, deviceScoreEnd - deviceScoreStart);
+ if (status != DS_SUCCESS)
+ {
+ goto cleanup;
+ }
+ }
+ }
+ }
+
+ }
+ else if (deviceType == DS_DEVICE_NATIVE_CPU)
+ {
+ for (i = 0; i < profile->numDevices; i++)
+ {
+ if (profile->devices[i].type == DS_DEVICE_NATIVE_CPU)
+ {
+ deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
+ if (deviceScoreStart == NULL)
+ {
+ status = DS_PROFILE_FILE_ERROR;
+ goto cleanup;
+ }
+ deviceScoreStart += strlen(DS_TAG_SCORE);
+ deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
+ status = deserializer(profile->devices + i, (const unsigned char*)deviceScoreStart, deviceScoreEnd - deviceScoreStart);
+ if (status != DS_SUCCESS)
+ {
+ goto cleanup;
+ }
+ }
+ }
+ }
+
+ // skip over the current one to find the next device
+ currentPosition = dataEnd + strlen(DS_TAG_DEVICE_END);
+ }
+ }
+cleanup:
+ if (contentStart != NULL) free(contentStart);
+ return status;
+}
+
+static ds_status getNumDeviceWithEmptyScore(ds_profile* profile, unsigned int* num)
+{
+ unsigned int i;
+ if (profile == NULL || num == NULL) return DS_MEMORY_ERROR;
+ *num = 0;
+ for (i = 0; i < profile->numDevices; i++)
+ {
+ if (profile->devices[i].score == NULL)
+ {
+ *num++;
+ }
+ }
+ return DS_SUCCESS;
+}
+
+#endif
+
+/* vim:set shiftwidth=4 softtabstop=4 expandtab: */
diff --git a/sc/source/core/opencl/openclwrapper.cxx b/sc/source/core/opencl/openclwrapper.cxx
index 9a056aa..efe52a4 100644
--- a/sc/source/core/opencl/openclwrapper.cxx
+++ b/sc/source/core/opencl/openclwrapper.cxx
@@ -19,6 +19,7 @@
#include "sal/config.h"
#include <osl/file.hxx>
+#include "opencl_device.hxx"
#include <stdio.h>
#include <stdlib.h>
@@ -898,23 +899,14 @@ bool switchOpenclDevice(const OUString* pDevice, bool bAutoSelect)
if(!pDeviceId || bAutoSelect)
{
- size_t nComputeUnits = 0;
- // clever algorithm
- const std::vector<OpenclPlatformInfo>& rPlatform = fillOpenCLInfo();
- for(std::vector<OpenclPlatformInfo>::const_iterator it =
- rPlatform.begin(), itEnd = rPlatform.end(); it != itEnd; ++it)
- {
- for(std::vector<OpenclDeviceInfo>::const_iterator itr =
- it->maDevices.begin(), itrEnd = it->maDevices.end();
- itr != itrEnd; ++itr)
- {
- if(itr->mnComputeUnits > nComputeUnits)
- {
- pDeviceId = reinterpret_cast<cl_device_id>(itr->device);
- nComputeUnits = itr->mnComputeUnits;
- }
- }
- }
+ OUString url("${$BRAND_BASE_DIR/" LIBO_ETC_FOLDER "/" SAL_CONFIGFILE("bootstrap") ":UserInstallation}/cache/");
+ rtl::Bootstrap::expandMacros(url);
+ OUString path;
+ osl::FileBase::getSystemPathFromFileURL(url,path);
+ OString dsFileName = rtl::OUStringToOString(path, RTL_TEXTENCODING_UTF8);
+ ds_device pSelectedDevice = sc::OpenCLDevice::getDeviceSelection(dsFileName.getStr());
+ pDeviceId = pSelectedDevice.oclDeviceID;
+
}
if(OpenclDevice::gpuEnv.mpDevID == pDeviceId)
More information about the Libreoffice-commits
mailing list