1 /*
2 * Copyright (C) 2018-2021 Intel Corporation
3 *
4 * SPDX-License-Identifier: MIT
5 *
6 */
7
8 #include "offline_compiler_tests.h"
9
10 #include "shared/source/compiler_interface/intermediate_representations.h"
11 #include "shared/source/compiler_interface/oclc_extensions.h"
12 #include "shared/source/debug_settings/debug_settings_manager.h"
13 #include "shared/source/device_binary_format/elf/elf_decoder.h"
14 #include "shared/source/device_binary_format/elf/ocl_elf.h"
15 #include "shared/source/helpers/file_io.h"
16 #include "shared/source/helpers/hw_info.h"
17 #include "shared/test/common/helpers/debug_manager_state_restore.h"
18 #include "shared/test/common/mocks/mock_compilers.h"
19 #include "shared/test/unit_test/device_binary_format/zebin_tests.h"
20
21 #include "compiler_options.h"
22 #include "environment.h"
23 #include "gmock/gmock.h"
24 #include "hw_cmds.h"
25 #include "mock/mock_argument_helper.h"
26 #include "mock/mock_offline_compiler.h"
27
28 #include <algorithm>
29 #include <array>
30 #include <fstream>
31
32 extern Environment *gEnvironment;
33
34 namespace NEO {
35
createFileWithArgs(const std::vector<std::string> & singleArgs,int numOfBuild)36 void MultiCommandTests::createFileWithArgs(const std::vector<std::string> &singleArgs, int numOfBuild) {
37 std::ofstream myfile(nameOfFileWithArgs);
38 if (myfile.is_open()) {
39 for (int i = 0; i < numOfBuild; i++) {
40 for (auto singleArg : singleArgs)
41 myfile << singleArg + " ";
42 myfile << std::endl;
43 }
44 myfile.close();
45 } else
46 printf("Unable to open file\n");
47 }
48
deleteFileWithArgs()49 void MultiCommandTests::deleteFileWithArgs() {
50 if (remove(nameOfFileWithArgs.c_str()) != 0)
51 perror("Error deleting file");
52 }
53
deleteOutFileList()54 void MultiCommandTests::deleteOutFileList() {
55 if (remove(outFileList.c_str()) != 0)
56 perror("Error deleting file");
57 }
58
getCompilerOutputFileName(const std::string & fileName,const std::string & type)59 std::string getCompilerOutputFileName(const std::string &fileName, const std::string &type) {
60 std::string fName(fileName);
61 fName.append("_");
62 fName.append(gEnvironment->familyNameWithType);
63 fName.append(".");
64 fName.append(type);
65 return fName;
66 }
67
compilerOutputExists(const std::string & fileName,const std::string & type)68 bool compilerOutputExists(const std::string &fileName, const std::string &type) {
69 return fileExists(getCompilerOutputFileName(fileName, type));
70 }
71
compilerOutputRemove(const std::string & fileName,const std::string & type)72 void compilerOutputRemove(const std::string &fileName, const std::string &type) {
73 std::remove(getCompilerOutputFileName(fileName, type).c_str());
74 }
75
76 template <typename SectionHeaders>
isAnyIrSectionDefined(const SectionHeaders & sectionHeaders)77 bool isAnyIrSectionDefined(const SectionHeaders §ionHeaders) {
78 const auto isIrSection = [](const auto §ion) {
79 return section.header && (section.header->type == Elf::SHT_OPENCL_SPIRV || section.header->type == Elf::SHT_OPENCL_LLVM_BINARY);
80 };
81
82 return std::any_of(sectionHeaders.begin(), sectionHeaders.end(), isIrSection);
83 }
84
TEST_F(MultiCommandTests,WhenBuildingMultiCommandThenSuccessIsReturned)85 TEST_F(MultiCommandTests, WhenBuildingMultiCommandThenSuccessIsReturned) {
86 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
87 std::vector<std::string> argv = {
88 "ocloc",
89 "multi",
90 nameOfFileWithArgs.c_str(),
91 "-q",
92 };
93
94 std::vector<std::string> singleArgs = {
95 "-file",
96 "test_files/copybuffer.cl",
97 "-device",
98 gEnvironment->devicePrefix.c_str()};
99
100 int numOfBuild = 4;
101 createFileWithArgs(singleArgs, numOfBuild);
102
103 auto pMultiCommand = std::unique_ptr<MultiCommand>(MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get()));
104
105 EXPECT_NE(nullptr, pMultiCommand);
106 EXPECT_EQ(CL_SUCCESS, retVal);
107
108 deleteFileWithArgs();
109 }
110
TEST_F(MultiCommandTests,GivenOutputFileWhenBuildingMultiCommandThenSuccessIsReturned)111 TEST_F(MultiCommandTests, GivenOutputFileWhenBuildingMultiCommandThenSuccessIsReturned) {
112 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
113 std::vector<std::string> argv = {
114 "ocloc",
115 "multi",
116 nameOfFileWithArgs.c_str(),
117 "-q",
118 };
119
120 std::vector<std::string> singleArgs = {
121 "-file",
122 "test_files/copybuffer.cl",
123 "-device",
124 gEnvironment->devicePrefix.c_str()};
125
126 int numOfBuild = 4;
127 createFileWithArgs(singleArgs, numOfBuild);
128
129 auto pMultiCommand = std::unique_ptr<MultiCommand>(MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get()));
130
131 EXPECT_NE(nullptr, pMultiCommand);
132 EXPECT_EQ(CL_SUCCESS, retVal);
133
134 for (int i = 0; i < numOfBuild; i++) {
135 std::string outFileName = pMultiCommand->outDirForBuilds + "/build_no_" + std::to_string(i + 1);
136 EXPECT_TRUE(compilerOutputExists(outFileName, "bc") || compilerOutputExists(outFileName, "spv"));
137 EXPECT_TRUE(compilerOutputExists(outFileName, "gen"));
138 EXPECT_TRUE(compilerOutputExists(outFileName, "bin"));
139 }
140
141 deleteFileWithArgs();
142 }
143
TEST_F(MultiCommandTests,GivenSpecifiedOutputDirWhenBuildingMultiCommandThenSuccessIsReturned)144 TEST_F(MultiCommandTests, GivenSpecifiedOutputDirWhenBuildingMultiCommandThenSuccessIsReturned) {
145 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
146 std::vector<std::string> argv = {
147 "ocloc",
148 "multi",
149 nameOfFileWithArgs.c_str(),
150 "-q",
151 };
152
153 std::vector<std::string> singleArgs = {
154 "-file",
155 "test_files/copybuffer.cl",
156 "-device",
157 gEnvironment->devicePrefix.c_str(),
158 "-out_dir",
159 "offline_compiler_test"};
160
161 int numOfBuild = 4;
162 createFileWithArgs(singleArgs, numOfBuild);
163
164 pMultiCommand = MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get());
165
166 EXPECT_NE(nullptr, pMultiCommand);
167 EXPECT_EQ(CL_SUCCESS, retVal);
168
169 for (int i = 0; i < numOfBuild; i++) {
170 std::string outFileName = "offline_compiler_test/build_no_" + std::to_string(i + 1);
171 EXPECT_TRUE(compilerOutputExists(outFileName, "bc") || compilerOutputExists(outFileName, "spv"));
172 EXPECT_TRUE(compilerOutputExists(outFileName, "gen"));
173 EXPECT_TRUE(compilerOutputExists(outFileName, "bin"));
174 }
175
176 deleteFileWithArgs();
177 delete pMultiCommand;
178 }
179
TEST_F(MultiCommandTests,GivenSpecifiedOutputDirWithProductConfigValueWhenBuildingMultiCommandThenSuccessIsReturned)180 TEST_F(MultiCommandTests, GivenSpecifiedOutputDirWithProductConfigValueWhenBuildingMultiCommandThenSuccessIsReturned) {
181 auto allEnabledDeviceConfigs = oclocArgHelperWithoutInput->getAllSupportedDeviceConfigs();
182 if (allEnabledDeviceConfigs.empty()) {
183 GTEST_SKIP();
184 }
185 auto deviceMapConfig = allEnabledDeviceConfigs[0];
186 auto configStr = oclocArgHelperWithoutInput->parseProductConfigFromValue(deviceMapConfig.config);
187
188 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
189 std::vector<std::string> argv = {
190 "ocloc",
191 "multi",
192 nameOfFileWithArgs.c_str(),
193 "-q",
194 };
195
196 std::vector<std::string> singleArgs = {
197 "-file",
198 "test_files/copybuffer.cl",
199 "-device",
200 configStr,
201 "-out_dir",
202 "offline_compiler_test"};
203
204 int numOfBuild = 4;
205 createFileWithArgs(singleArgs, numOfBuild);
206
207 pMultiCommand = MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get());
208
209 EXPECT_NE(nullptr, pMultiCommand);
210 EXPECT_EQ(CL_SUCCESS, retVal);
211
212 for (int i = 0; i < numOfBuild; i++) {
213 std::string outFileName = "offline_compiler_test/build_no_" + std::to_string(i + 1);
214 EXPECT_TRUE(compilerOutputExists(outFileName, "bc") || compilerOutputExists(outFileName, "spv"));
215 EXPECT_TRUE(compilerOutputExists(outFileName, "gen"));
216 EXPECT_TRUE(compilerOutputExists(outFileName, "bin"));
217 }
218
219 deleteFileWithArgs();
220 delete pMultiCommand;
221 }
222
TEST_F(MultiCommandTests,GivenMissingTextFileWithArgsWhenBuildingMultiCommandThenInvalidFileErrorIsReturned)223 TEST_F(MultiCommandTests, GivenMissingTextFileWithArgsWhenBuildingMultiCommandThenInvalidFileErrorIsReturned) {
224 nameOfFileWithArgs = "test_files/ImANotExistedComandFile.txt";
225 std::vector<std::string> argv = {
226 "ocloc",
227 "multi",
228 "test_files/ImANaughtyFile.txt",
229 "-q",
230 };
231
232 testing::internal::CaptureStdout();
233 auto pMultiCommand = std::unique_ptr<MultiCommand>(MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get()));
234 std::string output = testing::internal::GetCapturedStdout();
235
236 EXPECT_STRNE(output.c_str(), "");
237 EXPECT_EQ(nullptr, pMultiCommand);
238 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_FILE, retVal);
239 DebugManager.flags.PrintDebugMessages.set(false);
240 }
TEST_F(MultiCommandTests,GivenLackOfClFileWhenBuildingMultiCommandThenInvalidFileErrorIsReturned)241 TEST_F(MultiCommandTests, GivenLackOfClFileWhenBuildingMultiCommandThenInvalidFileErrorIsReturned) {
242 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
243 std::vector<std::string> argv = {
244 "ocloc",
245 "multi",
246 nameOfFileWithArgs.c_str(),
247 "-q",
248 };
249
250 std::vector<std::string> singleArgs = {
251 "-file",
252 "test_files/ImANaughtyFile.cl",
253 "-device",
254 gEnvironment->devicePrefix.c_str()};
255
256 int numOfBuild = 4;
257 createFileWithArgs(singleArgs, numOfBuild);
258 testing::internal::CaptureStdout();
259 auto pMultiCommand = std::unique_ptr<MultiCommand>(MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get()));
260 std::string output = testing::internal::GetCapturedStdout();
261
262 EXPECT_EQ(nullptr, pMultiCommand);
263 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_FILE, retVal);
264 DebugManager.flags.PrintDebugMessages.set(false);
265
266 deleteFileWithArgs();
267 }
TEST_F(MultiCommandTests,GivenOutputFileListFlagWhenBuildingMultiCommandThenSuccessIsReturned)268 TEST_F(MultiCommandTests, GivenOutputFileListFlagWhenBuildingMultiCommandThenSuccessIsReturned) {
269 nameOfFileWithArgs = "test_files/ImAMulitiComandMinimalGoodFile.txt";
270 std::vector<std::string> argv = {
271 "ocloc",
272 "multi",
273 nameOfFileWithArgs.c_str(),
274 "-q",
275 "-output_file_list",
276 "outFileList.txt",
277 };
278
279 std::vector<std::string> singleArgs = {
280 "-file",
281 "test_files/copybuffer.cl",
282 "-device",
283 gEnvironment->devicePrefix.c_str()};
284
285 int numOfBuild = 4;
286 createFileWithArgs(singleArgs, numOfBuild);
287
288 pMultiCommand = MultiCommand::create(argv, retVal, oclocArgHelperWithoutInput.get());
289
290 EXPECT_NE(nullptr, pMultiCommand);
291 EXPECT_EQ(CL_SUCCESS, retVal);
292 outFileList = pMultiCommand->outputFileList;
293 EXPECT_TRUE(fileExists(outFileList));
294
295 for (int i = 0; i < numOfBuild; i++) {
296 std::string outFileName = pMultiCommand->outDirForBuilds + "/build_no_" + std::to_string(i + 1);
297 EXPECT_TRUE(compilerOutputExists(outFileName, "bc") || compilerOutputExists(outFileName, "spv"));
298 EXPECT_TRUE(compilerOutputExists(outFileName, "gen"));
299 EXPECT_TRUE(compilerOutputExists(outFileName, "bin"));
300 }
301
302 deleteFileWithArgs();
303 deleteOutFileList();
304 delete pMultiCommand;
305 }
306
TEST_F(OfflineCompilerTests,GivenHelpOptionOnQueryThenSuccessIsReturned)307 TEST_F(OfflineCompilerTests, GivenHelpOptionOnQueryThenSuccessIsReturned) {
308 std::vector<std::string> argv = {
309 "ocloc",
310 "query",
311 "--help"};
312
313 testing::internal::CaptureStdout();
314 int retVal = OfflineCompiler::query(argv.size(), argv, oclocArgHelperWithoutInput.get());
315 std::string output = testing::internal::GetCapturedStdout();
316
317 EXPECT_STREQ(OfflineCompiler::queryHelp.data(), output.c_str());
318 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
319 }
320
TEST_F(OfflineCompilerTests,GivenArgsWhenQueryIsCalledThenSuccessIsReturned)321 TEST_F(OfflineCompilerTests, GivenArgsWhenQueryIsCalledThenSuccessIsReturned) {
322 std::vector<std::string> argv = {
323 "ocloc",
324 "query",
325 "NEO_REVISION"};
326
327 int retVal = OfflineCompiler::query(argv.size(), argv, oclocArgHelperWithoutInput.get());
328
329 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
330 }
331
TEST_F(OfflineCompilerTests,GivenArgsWhenOfflineCompilerIsCreatedThenSuccessIsReturned)332 TEST_F(OfflineCompilerTests, GivenArgsWhenOfflineCompilerIsCreatedThenSuccessIsReturned) {
333 std::vector<std::string> argv = {
334 "ocloc",
335 "-file",
336 "test_files/copybuffer.cl",
337 "-device",
338 gEnvironment->devicePrefix.c_str()};
339
340 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
341
342 EXPECT_NE(nullptr, pOfflineCompiler);
343 EXPECT_EQ(CL_SUCCESS, retVal);
344
345 delete pOfflineCompiler;
346 }
347
TEST_F(OfflineCompilerTests,givenProperDeviceIdHexAsDeviceArgumentThenSuccessIsReturned)348 TEST_F(OfflineCompilerTests, givenProperDeviceIdHexAsDeviceArgumentThenSuccessIsReturned) {
349 std::map<std::string, std::string> files;
350 std::unique_ptr<MockOclocArgHelper> argHelper = std::make_unique<MockOclocArgHelper>(files);
351
352 if (argHelper->deviceProductTable.size() == 1 && argHelper->deviceProductTable[0].deviceId == 0) {
353 GTEST_SKIP();
354 }
355 std::stringstream deviceString, productString;
356 deviceString << "0x" << std::hex << argHelper->deviceProductTable[0].deviceId;
357 productString << argHelper->deviceProductTable[0].product;
358
359 std::vector<std::string> argv = {
360 "ocloc",
361 "-file",
362 "test_files/copybuffer.cl",
363 "-device",
364 deviceString.str()};
365 testing::internal::CaptureStdout();
366 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
367 auto output = testing::internal::GetCapturedStdout();
368 std::stringstream resString;
369 resString << "Auto-detected target based on " << deviceString.str() << " device id: " << productString.str() << "\n";
370
371 EXPECT_NE(nullptr, pOfflineCompiler);
372 EXPECT_STREQ(output.c_str(), resString.str().c_str());
373 EXPECT_EQ(CL_SUCCESS, retVal);
374
375 delete pOfflineCompiler;
376 }
377
TEST_F(OfflineCompilerTests,givenIncorrectDeviceIdHexThenInvalidDeviceIsReturned)378 TEST_F(OfflineCompilerTests, givenIncorrectDeviceIdHexThenInvalidDeviceIsReturned) {
379 std::vector<std::string> argv = {
380 "ocloc",
381 "-file",
382 "test_files/copybuffer.cl",
383 "-device",
384 "0x0"};
385 testing::internal::CaptureStdout();
386 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
387 auto output = testing::internal::GetCapturedStdout();
388 EXPECT_EQ(nullptr, pOfflineCompiler);
389 EXPECT_STREQ(output.c_str(), "Could not determine target based on device id: 0x0\nError: Cannot get HW Info for device 0x0.\n");
390 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
391 }
392
TEST_F(OfflineCompilerTests,givenDeviceNumerationWithMissingRevisionValueWhenInvalidPatternIsPassedThenInvalidDeviceIsReturned)393 TEST_F(OfflineCompilerTests, givenDeviceNumerationWithMissingRevisionValueWhenInvalidPatternIsPassedThenInvalidDeviceIsReturned) {
394 std::vector<std::string> argv = {
395 "ocloc",
396 "-file",
397 "test_files/copybuffer.cl",
398 "-device",
399 "9.1."};
400 testing::internal::CaptureStdout();
401 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
402 auto output = testing::internal::GetCapturedStdout();
403 EXPECT_EQ(nullptr, pOfflineCompiler);
404 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: 9.1.\nError: Cannot get HW Info for device 9.1..\n");
405 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
406 }
407
TEST_F(OfflineCompilerTests,givenDeviceNumerationWithInvalidPatternThenInvalidDeviceIsReturned)408 TEST_F(OfflineCompilerTests, givenDeviceNumerationWithInvalidPatternThenInvalidDeviceIsReturned) {
409 std::vector<std::string> argv = {
410 "ocloc",
411 "-file",
412 "test_files/copybuffer.cl",
413 "-device",
414 "9.1.."};
415 testing::internal::CaptureStdout();
416 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
417 auto output = testing::internal::GetCapturedStdout();
418 EXPECT_EQ(nullptr, pOfflineCompiler);
419 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: 9.1..\nError: Cannot get HW Info for device 9.1...\n");
420 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
421 }
422
TEST_F(OfflineCompilerTests,givenDeviceNumerationWithMissingMajorValueWhenInvalidPatternIsPassedThenInvalidDeviceIsReturned)423 TEST_F(OfflineCompilerTests, givenDeviceNumerationWithMissingMajorValueWhenInvalidPatternIsPassedThenInvalidDeviceIsReturned) {
424 std::vector<std::string> argv = {
425 "ocloc",
426 "-file",
427 "test_files/copybuffer.cl",
428 "-device",
429 ".1.2"};
430 testing::internal::CaptureStdout();
431 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
432 auto output = testing::internal::GetCapturedStdout();
433 EXPECT_EQ(nullptr, pOfflineCompiler);
434 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: .1.2\nError: Cannot get HW Info for device .1.2.\n");
435 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
436 }
437
TEST_F(OfflineCompilerTests,givenDeviceNumerationWhenInvalidRevisionValueIsPassedThenInvalidDeviceIsReturned)438 TEST_F(OfflineCompilerTests, givenDeviceNumerationWhenInvalidRevisionValueIsPassedThenInvalidDeviceIsReturned) {
439 std::vector<std::string> argv = {
440 "ocloc",
441 "-file",
442 "test_files/copybuffer.cl",
443 "-device",
444 "9.0.a"};
445 testing::internal::CaptureStdout();
446 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
447 auto output = testing::internal::GetCapturedStdout();
448 EXPECT_EQ(nullptr, pOfflineCompiler);
449 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: 9.0.a\nError: Cannot get HW Info for device 9.0.a.\n");
450 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
451 }
452
TEST_F(OfflineCompilerTests,givenDeviceNumerationWhenInvalidMinorValueIsPassedThenInvalidDeviceIsReturned)453 TEST_F(OfflineCompilerTests, givenDeviceNumerationWhenInvalidMinorValueIsPassedThenInvalidDeviceIsReturned) {
454 std::vector<std::string> argv = {
455 "ocloc",
456 "-file",
457 "test_files/copybuffer.cl",
458 "-device",
459 "9.a"};
460 testing::internal::CaptureStdout();
461 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
462 auto output = testing::internal::GetCapturedStdout();
463 EXPECT_EQ(nullptr, pOfflineCompiler);
464 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: 9.a\nError: Cannot get HW Info for device 9.a.\n");
465 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
466 }
467
TEST_F(OfflineCompilerTests,givenDeviceNumerationWhenPassedValuesAreOutOfRangeThenInvalidDeviceIsReturned)468 TEST_F(OfflineCompilerTests, givenDeviceNumerationWhenPassedValuesAreOutOfRangeThenInvalidDeviceIsReturned) {
469 std::vector<std::string> argv = {
470 "ocloc",
471 "-file",
472 "test_files/copybuffer.cl",
473 "-device",
474 "256.350"};
475 testing::internal::CaptureStdout();
476 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
477 auto output = testing::internal::GetCapturedStdout();
478 EXPECT_EQ(nullptr, pOfflineCompiler);
479 EXPECT_STREQ(output.c_str(), "Could not determine target based on product config: 256.350\nError: Cannot get HW Info for device 256.350.\n");
480 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
481 }
482
TEST_F(OfflineCompilerTests,givenIncorrectDeviceIdWithIncorrectHexPatternThenInvalidDeviceIsReturned)483 TEST_F(OfflineCompilerTests, givenIncorrectDeviceIdWithIncorrectHexPatternThenInvalidDeviceIsReturned) {
484 std::vector<std::string> argv = {
485 "ocloc",
486 "-file",
487 "test_files/copybuffer.cl",
488 "-device",
489 "0xnonexist"};
490 testing::internal::CaptureStdout();
491 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
492 auto output = testing::internal::GetCapturedStdout();
493 EXPECT_EQ(nullptr, pOfflineCompiler);
494 EXPECT_STREQ(output.c_str(), "Error: Cannot get HW Info for device 0xnonexist.\n");
495 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
496 }
497
TEST_F(OfflineCompilerTests,givenDebugOptionThenInternalOptionShouldContainKernelDebugEnable)498 TEST_F(OfflineCompilerTests, givenDebugOptionThenInternalOptionShouldContainKernelDebugEnable) {
499 if (gEnvironment->devicePrefix == "bdw") {
500 GTEST_SKIP();
501 }
502
503 std::vector<std::string> argv = {
504 "ocloc",
505 "-options",
506 "-g",
507 "-file",
508 "test_files/copybuffer.cl",
509 "-device",
510 gEnvironment->devicePrefix.c_str()};
511
512 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
513 ASSERT_NE(nullptr, mockOfflineCompiler);
514 mockOfflineCompiler->initialize(argv.size(), argv);
515
516 std::string internalOptions = mockOfflineCompiler->internalOptions;
517 EXPECT_THAT(internalOptions, ::testing::HasSubstr("-cl-kernel-debug-enable"));
518 }
519
TEST_F(OfflineCompilerTests,givenDashGInBiggerOptionStringWhenInitializingThenInternalOptionsShouldNotContainKernelDebugEnable)520 TEST_F(OfflineCompilerTests, givenDashGInBiggerOptionStringWhenInitializingThenInternalOptionsShouldNotContainKernelDebugEnable) {
521
522 std::vector<std::string> argv = {
523 "ocloc",
524 "-options",
525 "-gNotRealDashGOption",
526 "-file",
527 "test_files/copybuffer.cl",
528 "-device",
529 gEnvironment->devicePrefix.c_str()};
530
531 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
532 ASSERT_NE(nullptr, mockOfflineCompiler);
533 mockOfflineCompiler->initialize(argv.size(), argv);
534
535 std::string internalOptions = mockOfflineCompiler->internalOptions;
536 EXPECT_THAT(internalOptions, ::testing::Not(::testing::HasSubstr("-cl-kernel-debug-enable")));
537 }
538
TEST_F(OfflineCompilerTests,givenExcludeIrFromZebinInternalOptionWhenInitIsPerformedThenIrExcludeFlagsShouldBeUnified)539 TEST_F(OfflineCompilerTests, givenExcludeIrFromZebinInternalOptionWhenInitIsPerformedThenIrExcludeFlagsShouldBeUnified) {
540 std::vector<std::string> argv = {
541 "ocloc",
542 "-file",
543 "test_files/copybuffer.cl",
544 "-internal_options",
545 "-ze-allow-zebin -ze-exclude-ir-from-zebin",
546 "-device",
547 gEnvironment->devicePrefix.c_str()};
548
549 MockOfflineCompiler mockOfflineCompiler{};
550 mockOfflineCompiler.initialize(argv.size(), argv);
551
552 EXPECT_TRUE(mockOfflineCompiler.excludeIr);
553 }
554
TEST_F(OfflineCompilerTests,givenExcludeIrArgumentWhenInitIsPerformedThenIrExcludeFlagsShouldBeUnified)555 TEST_F(OfflineCompilerTests, givenExcludeIrArgumentWhenInitIsPerformedThenIrExcludeFlagsShouldBeUnified) {
556 std::vector<std::string> argv = {
557 "ocloc",
558 "-file",
559 "test_files/copybuffer.cl",
560 "-exclude_ir",
561 "-internal_options",
562 "-ze-allow-zebin",
563 "-device",
564 gEnvironment->devicePrefix.c_str()};
565
566 MockOfflineCompiler mockOfflineCompiler{};
567 mockOfflineCompiler.initialize(argv.size(), argv);
568
569 const auto expectedInternalOption{"-ze-exclude-ir-from-zebin"};
570 const auto excludeIrFromZebinEnabled{mockOfflineCompiler.internalOptions.find(expectedInternalOption) != std::string::npos};
571 EXPECT_TRUE(excludeIrFromZebinEnabled);
572 }
573
TEST_F(OfflineCompilerTests,givenExcludeIrArgumentWhenCompilingKernelThenIrShouldBeExcluded)574 TEST_F(OfflineCompilerTests, givenExcludeIrArgumentWhenCompilingKernelThenIrShouldBeExcluded) {
575 std::vector<std::string> argv = {
576 "ocloc",
577 "-file",
578 "test_files/copybuffer.cl",
579 "-exclude_ir",
580 "-device",
581 gEnvironment->devicePrefix.c_str()};
582
583 MockOfflineCompiler mockOfflineCompiler{};
584 mockOfflineCompiler.initialize(argv.size(), argv);
585
586 const auto buildResult{mockOfflineCompiler.build()};
587 ASSERT_EQ(OfflineCompiler::SUCCESS, buildResult);
588
589 std::string errorReason{};
590 std::string warning{};
591
592 const auto elf{Elf::decodeElf(mockOfflineCompiler.elfBinary, errorReason, warning)};
593 ASSERT_TRUE(errorReason.empty());
594 ASSERT_TRUE(warning.empty());
595
596 EXPECT_FALSE(isAnyIrSectionDefined(elf.sectionHeaders));
597 }
598
TEST_F(OfflineCompilerTests,givenLackOfExcludeIrArgumentWhenCompilingKernelThenIrShouldBeIncluded)599 TEST_F(OfflineCompilerTests, givenLackOfExcludeIrArgumentWhenCompilingKernelThenIrShouldBeIncluded) {
600 std::vector<std::string> argv = {
601 "ocloc",
602 "-file",
603 "test_files/copybuffer.cl",
604 "-device",
605 gEnvironment->devicePrefix.c_str()};
606
607 MockOfflineCompiler mockOfflineCompiler{};
608 mockOfflineCompiler.initialize(argv.size(), argv);
609
610 const auto buildResult{mockOfflineCompiler.build()};
611 ASSERT_EQ(OfflineCompiler::SUCCESS, buildResult);
612
613 std::string errorReason{};
614 std::string warning{};
615
616 const auto elf{Elf::decodeElf(mockOfflineCompiler.elfBinary, errorReason, warning)};
617 ASSERT_TRUE(errorReason.empty());
618 ASSERT_TRUE(warning.empty());
619
620 EXPECT_TRUE(isAnyIrSectionDefined(elf.sectionHeaders));
621 }
622
TEST_F(OfflineCompilerTests,givenVariousClStdValuesWhenCompilingSourceThenCorrectExtensionsArePassed)623 TEST_F(OfflineCompilerTests, givenVariousClStdValuesWhenCompilingSourceThenCorrectExtensionsArePassed) {
624 std::string clStdOptionValues[] = {"", "-cl-std=CL1.2", "-cl-std=CL2.0", "-cl-std=CL3.0"};
625
626 for (auto &clStdOptionValue : clStdOptionValues) {
627 std::vector<std::string> argv = {
628 "ocloc",
629 "-file",
630 "test_files/copybuffer.cl",
631 "-device",
632 gEnvironment->devicePrefix.c_str()};
633
634 if (!clStdOptionValue.empty()) {
635 argv.push_back("-options");
636 argv.push_back(clStdOptionValue);
637 }
638
639 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
640 ASSERT_NE(nullptr, mockOfflineCompiler);
641 mockOfflineCompiler->initialize(argv.size(), argv);
642
643 std::string internalOptions = mockOfflineCompiler->internalOptions;
644 std::string oclVersionOption = getOclVersionCompilerInternalOption(DEFAULT_PLATFORM::hwInfo.capabilityTable.clVersionSupport);
645 EXPECT_THAT(internalOptions, ::testing::HasSubstr(oclVersionOption));
646
647 if (clStdOptionValue == "-cl-std=CL2.0") {
648 auto expectedRegex = std::string{"cl_khr_3d_image_writes"};
649 if (DEFAULT_PLATFORM::hwInfo.capabilityTable.supportsImages) {
650 expectedRegex += ".+" + std::string{"cl_khr_3d_image_writes"};
651 }
652 EXPECT_THAT(internalOptions, ::testing::ContainsRegex(expectedRegex));
653 }
654
655 OpenClCFeaturesContainer openclCFeatures;
656 getOpenclCFeaturesList(DEFAULT_PLATFORM::hwInfo, openclCFeatures);
657 for (auto &feature : openclCFeatures) {
658 if (clStdOptionValue == "-cl-std=CL3.0") {
659 EXPECT_THAT(internalOptions, ::testing::HasSubstr(std::string{feature.name}));
660 } else {
661 EXPECT_THAT(internalOptions, ::testing::Not(::testing::HasSubstr(std::string{feature.name})));
662 }
663 }
664
665 if (DEFAULT_PLATFORM::hwInfo.capabilityTable.supportsImages) {
666 EXPECT_THAT(internalOptions, ::testing::HasSubstr(CompilerOptions::enableImageSupport.data()));
667 } else {
668 EXPECT_THAT(internalOptions, ::testing::Not(::testing::HasSubstr(CompilerOptions::enableImageSupport.data())));
669 }
670 }
671 }
672
TEST_F(OfflineCompilerTests,GivenArgsWhenBuildingThenBuildSucceeds)673 TEST_F(OfflineCompilerTests, GivenArgsWhenBuildingThenBuildSucceeds) {
674 std::vector<std::string> argv = {
675 "ocloc",
676 "-file",
677 "test_files/copybuffer.cl",
678 "-device",
679 gEnvironment->devicePrefix.c_str()};
680
681 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
682
683 EXPECT_NE(nullptr, pOfflineCompiler);
684 EXPECT_EQ(CL_SUCCESS, retVal);
685
686 testing::internal::CaptureStdout();
687 retVal = pOfflineCompiler->build();
688 std::string output = testing::internal::GetCapturedStdout();
689 EXPECT_EQ(CL_SUCCESS, retVal);
690 EXPECT_TRUE(compilerOutputExists("copybuffer", "bc") || compilerOutputExists("copybuffer", "spv"));
691 EXPECT_TRUE(compilerOutputExists("copybuffer", "gen"));
692 EXPECT_TRUE(compilerOutputExists("copybuffer", "bin"));
693
694 std::string buildLog = pOfflineCompiler->getBuildLog();
695 EXPECT_STREQ(buildLog.c_str(), "");
696
697 delete pOfflineCompiler;
698 }
699
TEST_F(OfflineCompilerTests,GivenArgsWhenBuildingWithDeviceConfigValueThenBuildSucceeds)700 TEST_F(OfflineCompilerTests, GivenArgsWhenBuildingWithDeviceConfigValueThenBuildSucceeds) {
701 auto allEnabledDeviceConfigs = oclocArgHelperWithoutInput->getAllSupportedDeviceConfigs();
702 if (allEnabledDeviceConfigs.empty()) {
703 return;
704 }
705 auto deviceMapConfig = allEnabledDeviceConfigs[0];
706 auto configString = oclocArgHelperWithoutInput->parseProductConfigFromValue(deviceMapConfig.config);
707
708 std::vector<std::string> argv = {
709 "ocloc",
710 "-file",
711 "test_files/copybuffer.cl",
712 "-device",
713 configString};
714
715 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
716
717 EXPECT_NE(nullptr, pOfflineCompiler);
718 EXPECT_EQ(CL_SUCCESS, retVal);
719
720 testing::internal::CaptureStdout();
721 retVal = pOfflineCompiler->build();
722 std::string output = testing::internal::GetCapturedStdout();
723 EXPECT_EQ(CL_SUCCESS, retVal);
724 EXPECT_TRUE(compilerOutputExists("copybuffer", "bc") || compilerOutputExists("copybuffer", "spv"));
725 EXPECT_TRUE(compilerOutputExists("copybuffer", "gen"));
726 EXPECT_TRUE(compilerOutputExists("copybuffer", "bin"));
727
728 std::string buildLog = pOfflineCompiler->getBuildLog();
729 EXPECT_STREQ(buildLog.c_str(), "");
730
731 delete pOfflineCompiler;
732 }
733
TEST_F(OfflineCompilerTests,GivenLlvmTextWhenBuildingThenBuildSucceeds)734 TEST_F(OfflineCompilerTests, GivenLlvmTextWhenBuildingThenBuildSucceeds) {
735 std::vector<std::string> argv = {
736 "ocloc",
737 "-file",
738 "test_files/copybuffer.cl",
739 "-device",
740 gEnvironment->devicePrefix.c_str(),
741 "-llvm_text"};
742
743 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
744
745 EXPECT_NE(nullptr, pOfflineCompiler);
746 EXPECT_EQ(CL_SUCCESS, retVal);
747
748 retVal = pOfflineCompiler->build();
749 EXPECT_EQ(CL_SUCCESS, retVal);
750 EXPECT_TRUE(compilerOutputExists("copybuffer", "ll"));
751 EXPECT_TRUE(compilerOutputExists("copybuffer", "gen"));
752 EXPECT_TRUE(compilerOutputExists("copybuffer", "bin"));
753
754 delete pOfflineCompiler;
755 }
756
TEST_F(OfflineCompilerTests,WhenFclNotNeededThenDontLoadIt)757 TEST_F(OfflineCompilerTests, WhenFclNotNeededThenDontLoadIt) {
758 std::vector<std::string> argv = {
759 "ocloc",
760 "-file",
761 "test_files/copybuffer.cl",
762 "-device",
763 gEnvironment->devicePrefix.c_str(),
764 "-spirv_input"};
765
766 MockOfflineCompiler offlineCompiler;
767 auto ret = offlineCompiler.initialize(argv.size(), argv);
768 EXPECT_EQ(0, ret);
769 EXPECT_EQ(nullptr, offlineCompiler.fclDeviceCtx);
770 EXPECT_NE(nullptr, offlineCompiler.igcDeviceCtx);
771 }
772
TEST_F(OfflineCompilerTests,WhenParsingBinToCharArrayThenCorrectResult)773 TEST_F(OfflineCompilerTests, WhenParsingBinToCharArrayThenCorrectResult) {
774 std::vector<std::string> argv = {
775 "ocloc",
776 "-file",
777 "test_files/copybuffer.cl",
778 "-device",
779 gEnvironment->devicePrefix.c_str()};
780
781 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
782 // clang-format off
783 uint8_t binary[] = {
784 0x02, 0x23, 0x3, 0x40, 0x56, 0x7, 0x80, 0x90, 0x1, 0x03,
785 0x34, 0x5, 0x60, 0x78, 0x9, 0x66, 0xff, 0x10, 0x10, 0x10,
786 0x02, 0x23, 0x3, 0x40, 0x56, 0x7, 0x80, 0x90, 0x1, 0x03,
787 0x34, 0x5, 0x60, 0x78, 0x9, 0x66, 0xff,
788 };
789 // clang-format on
790
791 std::string familyNameWithType = gEnvironment->familyNameWithType;
792 std::string fileName = "scheduler";
793 std::string retArray = pOfflineCompiler->parseBinAsCharArray(binary, sizeof(binary), fileName);
794 std::string target = "#include <cstddef>\n"
795 "#include <cstdint>\n\n"
796 "size_t SchedulerBinarySize_" +
797 familyNameWithType + " = 37;\n"
798 "uint32_t SchedulerBinary_" +
799 familyNameWithType + "[10] = {\n"
800 " 0x40032302, 0x90800756, 0x05340301, 0x66097860, 0x101010ff, 0x40032302, 0x90800756, 0x05340301, \n"
801 " 0x66097860, 0xff000000};\n\n"
802 "#include \"shared/source/built_ins/registry/built_ins_registry.h\"\n\n"
803 "namespace NEO {\n"
804 "static RegisterEmbeddedResource registerSchedulerBin(\n"
805 " \"" +
806 gEnvironment->familyNameWithType + "_0_scheduler.builtin_kernel.bin\",\n"
807 " (const char *)SchedulerBinary_" +
808 familyNameWithType + ",\n"
809 " SchedulerBinarySize_" +
810 familyNameWithType + ");\n"
811 "}\n";
812 EXPECT_EQ(retArray, target);
813
814 delete pOfflineCompiler;
815 }
TEST_F(OfflineCompilerTests,GivenCppFileWhenBuildingThenBuildSucceeds)816 TEST_F(OfflineCompilerTests, GivenCppFileWhenBuildingThenBuildSucceeds) {
817 std::vector<std::string> argv = {
818 "ocloc",
819 "-file",
820 "test_files/copybuffer.cl",
821 "-device",
822 gEnvironment->devicePrefix.c_str(),
823 "-cpp_file"};
824
825 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
826
827 EXPECT_NE(nullptr, pOfflineCompiler);
828 EXPECT_EQ(CL_SUCCESS, retVal);
829
830 retVal = pOfflineCompiler->build();
831 EXPECT_EQ(CL_SUCCESS, retVal);
832 EXPECT_TRUE(compilerOutputExists("copybuffer", "cpp"));
833 EXPECT_TRUE(compilerOutputExists("copybuffer", "bc") || compilerOutputExists("copybuffer", "spv"));
834 EXPECT_TRUE(compilerOutputExists("copybuffer", "gen"));
835 EXPECT_TRUE(compilerOutputExists("copybuffer", "bin"));
836
837 delete pOfflineCompiler;
838 }
TEST_F(OfflineCompilerTests,GivenOutputDirWhenBuildingThenBuildSucceeds)839 TEST_F(OfflineCompilerTests, GivenOutputDirWhenBuildingThenBuildSucceeds) {
840 std::vector<std::string> argv = {
841 "ocloc",
842 "-file",
843 "test_files/copybuffer.cl",
844 "-device",
845 gEnvironment->devicePrefix.c_str(),
846 "-out_dir",
847 "offline_compiler_test"};
848
849 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
850
851 EXPECT_NE(nullptr, pOfflineCompiler);
852 EXPECT_EQ(CL_SUCCESS, retVal);
853
854 retVal = pOfflineCompiler->build();
855 EXPECT_EQ(CL_SUCCESS, retVal);
856 EXPECT_TRUE(compilerOutputExists("offline_compiler_test/copybuffer", "bc") || compilerOutputExists("offline_compiler_test/copybuffer", "spv"));
857 EXPECT_TRUE(compilerOutputExists("offline_compiler_test/copybuffer", "gen"));
858 EXPECT_TRUE(compilerOutputExists("offline_compiler_test/copybuffer", "bin"));
859
860 delete pOfflineCompiler;
861 }
TEST_F(OfflineCompilerTests,GivenHelpOptionThenBuildDoesNotOccur)862 TEST_F(OfflineCompilerTests, GivenHelpOptionThenBuildDoesNotOccur) {
863 std::vector<std::string> argv = {
864 "ocloc",
865 "--help"};
866
867 testing::internal::CaptureStdout();
868 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
869 std::string output = testing::internal::GetCapturedStdout();
870 EXPECT_STRNE("", output.c_str());
871 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
872
873 delete pOfflineCompiler;
874 }
TEST_F(OfflineCompilerTests,GivenInvalidFileWhenBuildingThenInvalidFileErrorIsReturned)875 TEST_F(OfflineCompilerTests, GivenInvalidFileWhenBuildingThenInvalidFileErrorIsReturned) {
876 DebugManager.flags.PrintDebugMessages.set(true);
877 std::vector<std::string> argv = {
878 "ocloc",
879 "-file",
880 "test_files/ImANaughtyFile.cl",
881 "-device",
882 gEnvironment->devicePrefix.c_str()};
883
884 testing::internal::CaptureStdout();
885 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
886 std::string output = testing::internal::GetCapturedStdout();
887 EXPECT_STRNE(output.c_str(), "");
888 EXPECT_EQ(nullptr, pOfflineCompiler);
889 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_FILE, retVal);
890 DebugManager.flags.PrintDebugMessages.set(false);
891 delete pOfflineCompiler;
892 }
893
TEST_F(OfflineCompilerTests,GivenInvalidFlagWhenBuildingThenInvalidCommandLineErrorIsReturned)894 TEST_F(OfflineCompilerTests, GivenInvalidFlagWhenBuildingThenInvalidCommandLineErrorIsReturned) {
895 std::vector<std::string> argv = {
896 "ocloc",
897 "-n",
898 "test_files/ImANaughtyFile.cl",
899 "-device",
900 gEnvironment->devicePrefix.c_str()};
901
902 testing::internal::CaptureStdout();
903 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
904 std::string output = testing::internal::GetCapturedStdout();
905 EXPECT_STRNE(output.c_str(), "");
906 EXPECT_EQ(nullptr, pOfflineCompiler);
907 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_COMMAND_LINE, retVal);
908
909 delete pOfflineCompiler;
910 }
911
TEST_F(OfflineCompilerTests,GivenInvalidOptionsWhenBuildingThenInvalidCommandLineErrorIsReturned)912 TEST_F(OfflineCompilerTests, GivenInvalidOptionsWhenBuildingThenInvalidCommandLineErrorIsReturned) {
913 std::vector<std::string> argvA = {
914 "ocloc",
915 "-file",
916 };
917
918 testing::internal::CaptureStdout();
919 pOfflineCompiler = OfflineCompiler::create(argvA.size(), argvA, true, retVal, oclocArgHelperWithoutInput.get());
920 std::string output = testing::internal::GetCapturedStdout();
921 EXPECT_STRNE(output.c_str(), "");
922
923 EXPECT_EQ(nullptr, pOfflineCompiler);
924 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_COMMAND_LINE, retVal);
925
926 delete pOfflineCompiler;
927
928 std::vector<std::string> argvB = {
929 "ocloc",
930 "-file",
931 "test_files/ImANaughtyFile.cl",
932 "-device"};
933 testing::internal::CaptureStdout();
934 pOfflineCompiler = OfflineCompiler::create(argvB.size(), argvB, true, retVal, oclocArgHelperWithoutInput.get());
935 output = testing::internal::GetCapturedStdout();
936 EXPECT_STRNE(output.c_str(), "");
937 EXPECT_EQ(nullptr, pOfflineCompiler);
938 EXPECT_EQ(OfflineCompiler::ErrorCode::INVALID_COMMAND_LINE, retVal);
939
940 delete pOfflineCompiler;
941 }
942
TEST_F(OfflineCompilerTests,GivenNonexistantDeviceWhenCompilingThenInvalidDeviceErrorAndErrorMessageAreReturned)943 TEST_F(OfflineCompilerTests, GivenNonexistantDeviceWhenCompilingThenInvalidDeviceErrorAndErrorMessageAreReturned) {
944 std::vector<std::string> argv = {
945 "ocloc",
946 "-file",
947 "test_files/copybuffer.cl",
948 "-device",
949 "foobar"};
950
951 testing::internal::CaptureStdout();
952 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
953 std::string output = testing::internal::GetCapturedStdout();
954 EXPECT_STREQ(output.c_str(), "Error: Cannot get HW Info for device foobar.\n");
955 EXPECT_EQ(nullptr, pOfflineCompiler);
956 EXPECT_EQ(CL_INVALID_DEVICE, retVal);
957 }
958
TEST_F(OfflineCompilerTests,GivenInvalidKernelWhenBuildingThenBuildProgramFailureErrorIsReturned)959 TEST_F(OfflineCompilerTests, GivenInvalidKernelWhenBuildingThenBuildProgramFailureErrorIsReturned) {
960 std::vector<std::string> argv = {
961 "ocloc",
962 "-file",
963 "test_files/shouldfail.cl",
964 "-device",
965 gEnvironment->devicePrefix.c_str()};
966
967 pOfflineCompiler = OfflineCompiler::create(argv.size(), argv, true, retVal, oclocArgHelperWithoutInput.get());
968
969 EXPECT_NE(nullptr, pOfflineCompiler);
970 EXPECT_EQ(CL_SUCCESS, retVal);
971
972 gEnvironment->SetInputFileName("invalid_file_name");
973 testing::internal::CaptureStdout();
974
975 retVal = pOfflineCompiler->build();
976 EXPECT_EQ(CL_BUILD_PROGRAM_FAILURE, retVal);
977
978 std::string output = testing::internal::GetCapturedStdout();
979 EXPECT_STREQ(output.c_str(), "");
980
981 std::string buildLog = pOfflineCompiler->getBuildLog();
982 EXPECT_STRNE(buildLog.c_str(), "");
983
984 gEnvironment->SetInputFileName("copybuffer");
985
986 delete pOfflineCompiler;
987 }
988
TEST(OfflineCompilerTest,WhenParsingCmdLineThenOptionsAreReadCorrectly)989 TEST(OfflineCompilerTest, WhenParsingCmdLineThenOptionsAreReadCorrectly) {
990 std::vector<std::string> argv = {
991 "ocloc",
992 NEO::CompilerOptions::greaterThan4gbBuffersRequired.data()};
993
994 MockOfflineCompiler *mockOfflineCompiler = new MockOfflineCompiler();
995 ASSERT_NE(nullptr, mockOfflineCompiler);
996
997 testing::internal::CaptureStdout();
998 mockOfflineCompiler->parseCommandLine(argv.size(), argv);
999 std::string output = testing::internal::GetCapturedStdout();
1000
1001 std::string internalOptions = mockOfflineCompiler->internalOptions;
1002 size_t found = internalOptions.find(argv.begin()[1]);
1003 EXPECT_NE(std::string::npos, found);
1004
1005 delete mockOfflineCompiler;
1006 }
1007
TEST(OfflineCompilerTest,givenStatelessToStatefullOptimizationEnabledWhenDebugSettingsAreParsedThenOptimizationStringIsPresent)1008 TEST(OfflineCompilerTest, givenStatelessToStatefullOptimizationEnabledWhenDebugSettingsAreParsedThenOptimizationStringIsPresent) {
1009 DebugManagerStateRestore stateRestore;
1010 MockOfflineCompiler mockOfflineCompiler;
1011 DebugManager.flags.EnableStatelessToStatefulBufferOffsetOpt.set(1);
1012
1013 mockOfflineCompiler.parseDebugSettings();
1014
1015 std::string internalOptions = mockOfflineCompiler.internalOptions;
1016 size_t found = internalOptions.find(NEO::CompilerOptions::hasBufferOffsetArg.data());
1017 EXPECT_NE(std::string::npos, found);
1018 }
1019
TEST(OfflineCompilerTest,givenStatelessToStatefullOptimizationEnabledWhenDebugSettingsAreParsedThenOptimizationStringIsSetToDefault)1020 TEST(OfflineCompilerTest, givenStatelessToStatefullOptimizationEnabledWhenDebugSettingsAreParsedThenOptimizationStringIsSetToDefault) {
1021 DebugManagerStateRestore stateRestore;
1022 MockOfflineCompiler mockOfflineCompiler;
1023 DebugManager.flags.EnableStatelessToStatefulBufferOffsetOpt.set(-1);
1024
1025 mockOfflineCompiler.parseDebugSettings();
1026
1027 std::string internalOptions = mockOfflineCompiler.internalOptions;
1028 size_t found = internalOptions.find(NEO::CompilerOptions::hasBufferOffsetArg.data());
1029 EXPECT_NE(std::string::npos, found);
1030 }
1031
TEST(OfflineCompilerTest,GivenDelimitersWhenGettingStringThenParseIsCorrect)1032 TEST(OfflineCompilerTest, GivenDelimitersWhenGettingStringThenParseIsCorrect) {
1033 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1034 ASSERT_NE(nullptr, mockOfflineCompiler);
1035
1036 size_t srcSize = 0;
1037 auto ptrSrc = loadDataFromFile("test_files/copy_buffer_to_buffer.builtin_kernel", srcSize);
1038
1039 const std::string src = ptrSrc.get();
1040 ASSERT_EQ(srcSize, src.size());
1041
1042 // assert that pattern was found
1043 ASSERT_NE(std::string::npos, src.find("R\"===("));
1044 ASSERT_NE(std::string::npos, src.find(")===\""));
1045
1046 auto dst = mockOfflineCompiler->getStringWithinDelimiters(src);
1047
1048 size_t size = dst.size();
1049 char nullChar = '\0';
1050 EXPECT_EQ(nullChar, dst[size - 1]);
1051
1052 // expect that pattern was not found
1053 EXPECT_EQ(std::string::npos, dst.find("R\"===("));
1054 EXPECT_EQ(std::string::npos, dst.find(")===\""));
1055 }
1056
TEST(OfflineCompilerTest,WhenConvertingToPascalCaseThenResultIsCorrect)1057 TEST(OfflineCompilerTest, WhenConvertingToPascalCaseThenResultIsCorrect) {
1058 EXPECT_EQ(0, strcmp("AuxTranslation", convertToPascalCase("aux_translation").c_str()));
1059 EXPECT_EQ(0, strcmp("CopyBufferToBuffer", convertToPascalCase("copy_buffer_to_buffer").c_str()));
1060 EXPECT_EQ(0, strcmp("CopyBufferRect", convertToPascalCase("copy_buffer_rect").c_str()));
1061 EXPECT_EQ(0, strcmp("FillBuffer", convertToPascalCase("fill_buffer").c_str()));
1062 EXPECT_EQ(0, strcmp("CopyBufferToImage3d", convertToPascalCase("copy_buffer_to_image3d").c_str()));
1063 EXPECT_EQ(0, strcmp("CopyImage3dToBuffer", convertToPascalCase("copy_image3d_to_buffer").c_str()));
1064 EXPECT_EQ(0, strcmp("CopyImageToImage1d", convertToPascalCase("copy_image_to_image1d").c_str()));
1065 EXPECT_EQ(0, strcmp("CopyImageToImage2d", convertToPascalCase("copy_image_to_image2d").c_str()));
1066 EXPECT_EQ(0, strcmp("CopyImageToImage3d", convertToPascalCase("copy_image_to_image3d").c_str()));
1067 EXPECT_EQ(0, strcmp("FillImage1d", convertToPascalCase("fill_image1d").c_str()));
1068 EXPECT_EQ(0, strcmp("FillImage2d", convertToPascalCase("fill_image2d").c_str()));
1069 EXPECT_EQ(0, strcmp("FillImage3d", convertToPascalCase("fill_image3d").c_str()));
1070 EXPECT_EQ(0, strcmp("VmeBlockMotionEstimateIntel", convertToPascalCase("vme_block_motion_estimate_intel").c_str()));
1071 EXPECT_EQ(0, strcmp("VmeBlockAdvancedMotionEstimateCheckIntel", convertToPascalCase("vme_block_advanced_motion_estimate_check_intel").c_str()));
1072 EXPECT_EQ(0, strcmp("VmeBlockAdvancedMotionEstimateBidirectionalCheckIntel", convertToPascalCase("vme_block_advanced_motion_estimate_bidirectional_check_intel").c_str()));
1073 EXPECT_EQ(0, strcmp("Scheduler", convertToPascalCase("scheduler").c_str()));
1074 EXPECT_EQ(0, strcmp("", convertToPascalCase("").c_str()));
1075 }
1076
TEST(OfflineCompilerTest,GivenValidParamWhenGettingHardwareInfoThenSuccessIsReturned)1077 TEST(OfflineCompilerTest, GivenValidParamWhenGettingHardwareInfoThenSuccessIsReturned) {
1078 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1079 ASSERT_NE(nullptr, mockOfflineCompiler);
1080
1081 EXPECT_EQ(CL_INVALID_DEVICE, mockOfflineCompiler->initHardwareInfo("invalid"));
1082 EXPECT_EQ(PRODUCT_FAMILY::IGFX_UNKNOWN, mockOfflineCompiler->getHardwareInfo().platform.eProductFamily);
1083 EXPECT_EQ(CL_SUCCESS, mockOfflineCompiler->initHardwareInfo(gEnvironment->devicePrefix.c_str()));
1084 EXPECT_NE(PRODUCT_FAMILY::IGFX_UNKNOWN, mockOfflineCompiler->getHardwareInfo().platform.eProductFamily);
1085 }
1086
TEST(OfflineCompilerTest,GivenConfigValueWhichIsOutOfRangeWhenGettingHardwareInfoThenInvalidDeviceIsReturned)1087 TEST(OfflineCompilerTest, GivenConfigValueWhichIsOutOfRangeWhenGettingHardwareInfoThenInvalidDeviceIsReturned) {
1088 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1089 ASSERT_NE(nullptr, mockOfflineCompiler);
1090
1091 uint32_t value = 0xffffff + 1;
1092 std::stringstream inproperValue, resString;
1093 inproperValue << value;
1094
1095 testing::internal::CaptureStdout();
1096 EXPECT_EQ(CL_INVALID_DEVICE, mockOfflineCompiler->initHardwareInfo(inproperValue.str()));
1097
1098 resString << "Could not determine target based on product config: " << inproperValue.str() << "\n";
1099
1100 auto output = testing::internal::GetCapturedStdout();
1101 EXPECT_STREQ(output.c_str(), resString.str().c_str());
1102 }
1103
TEST(OfflineCompilerTest,WhenStoringBinaryThenStoredCorrectly)1104 TEST(OfflineCompilerTest, WhenStoringBinaryThenStoredCorrectly) {
1105 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1106 ASSERT_NE(nullptr, mockOfflineCompiler);
1107
1108 const char pSrcBinary[] = {0x01, 0x02, 0x03, 0x04, 0x05};
1109 const size_t srcBinarySize = sizeof(pSrcBinary);
1110 char *pDstBinary = new char[srcBinarySize];
1111 size_t dstBinarySize = srcBinarySize;
1112
1113 mockOfflineCompiler->storeBinary(pDstBinary, dstBinarySize, pSrcBinary, srcBinarySize);
1114 EXPECT_EQ(0, memcmp(pDstBinary, pSrcBinary, srcBinarySize));
1115
1116 delete[] pDstBinary;
1117 }
1118
TEST(OfflineCompilerTest,givenErrorStringsWithoutExtraNullCharactersWhenUpdatingBuildLogThenMessageIsCorrect)1119 TEST(OfflineCompilerTest, givenErrorStringsWithoutExtraNullCharactersWhenUpdatingBuildLogThenMessageIsCorrect) {
1120 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1121 ASSERT_NE(nullptr, mockOfflineCompiler);
1122
1123 std::string ErrorString = "Error: undefined variable";
1124 mockOfflineCompiler->updateBuildLog(ErrorString.c_str(), ErrorString.length());
1125 EXPECT_EQ(0, ErrorString.compare(mockOfflineCompiler->getBuildLog()));
1126
1127 std::string FinalString = "Build failure";
1128 mockOfflineCompiler->updateBuildLog(FinalString.c_str(), FinalString.length());
1129 EXPECT_EQ(0, (ErrorString + "\n" + FinalString).compare(mockOfflineCompiler->getBuildLog().c_str()));
1130 }
1131
TEST(OfflineCompilerTest,givenErrorStringsWithExtraNullCharactersWhenUpdatingBuildLogThenMessageIsCorrect)1132 TEST(OfflineCompilerTest, givenErrorStringsWithExtraNullCharactersWhenUpdatingBuildLogThenMessageIsCorrect) {
1133 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1134 ASSERT_NE(nullptr, mockOfflineCompiler);
1135
1136 std::array<char, sizeof("Error: undefined variable\0")> errorMessageArray = {"Error: undefined variable\0"};
1137 std::string expectedBuildLogString = "Error: undefined variable";
1138 EXPECT_EQ(errorMessageArray.size(), std::string("Error: undefined variable").length() + 2);
1139
1140 mockOfflineCompiler->updateBuildLog(errorMessageArray.data(), errorMessageArray.size());
1141 EXPECT_EQ(mockOfflineCompiler->getBuildLog(), expectedBuildLogString);
1142
1143 std::array<char, sizeof("Build failure\0")> additionalErrorMessageArray = {"Build failure\0"};
1144 expectedBuildLogString = "Error: undefined variable\n"
1145 "Build failure";
1146 EXPECT_EQ(additionalErrorMessageArray.size(), std::string("Build failure").length() + 2);
1147
1148 mockOfflineCompiler->updateBuildLog(additionalErrorMessageArray.data(), additionalErrorMessageArray.size());
1149 EXPECT_EQ(mockOfflineCompiler->getBuildLog(), expectedBuildLogString);
1150 }
1151
TEST(OfflineCompilerTest,GivenSourceCodeWhenBuildingThenSuccessIsReturned)1152 TEST(OfflineCompilerTest, GivenSourceCodeWhenBuildingThenSuccessIsReturned) {
1153 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1154 ASSERT_NE(nullptr, mockOfflineCompiler);
1155
1156 auto retVal = mockOfflineCompiler->buildSourceCode();
1157 EXPECT_EQ(CL_INVALID_PROGRAM, retVal);
1158
1159 std::vector<std::string> argv = {
1160 "ocloc",
1161 "-file",
1162 "test_files/copybuffer.cl",
1163 "-device",
1164 gEnvironment->devicePrefix.c_str()};
1165
1166 retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1167 EXPECT_EQ(CL_SUCCESS, retVal);
1168
1169 EXPECT_EQ(nullptr, mockOfflineCompiler->genBinary);
1170 EXPECT_EQ(0u, mockOfflineCompiler->genBinarySize);
1171
1172 retVal = mockOfflineCompiler->buildSourceCode();
1173 EXPECT_EQ(CL_SUCCESS, retVal);
1174
1175 EXPECT_NE(nullptr, mockOfflineCompiler->genBinary);
1176 EXPECT_NE(0u, mockOfflineCompiler->genBinarySize);
1177 }
1178
TEST(OfflineCompilerTest,givenSpvOnlyOptionPassedWhenCmdLineParsedThenGenerateOnlySpvFile)1179 TEST(OfflineCompilerTest, givenSpvOnlyOptionPassedWhenCmdLineParsedThenGenerateOnlySpvFile) {
1180 std::vector<std::string> argv = {
1181 "ocloc",
1182 "-file",
1183 "test_files/copybuffer.cl",
1184 "-output",
1185 "myOutputFileName",
1186 "-spv_only",
1187 "-device",
1188 gEnvironment->devicePrefix.c_str()};
1189
1190 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1191 ASSERT_NE(nullptr, mockOfflineCompiler);
1192
1193 auto retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1194 EXPECT_EQ(CL_SUCCESS, retVal);
1195
1196 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1197 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bin"));
1198 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "gen"));
1199
1200 retVal = mockOfflineCompiler->build();
1201 EXPECT_EQ(CL_SUCCESS, retVal);
1202
1203 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1204 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bin"));
1205 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "gen"));
1206
1207 compilerOutputRemove("myOutputFileName", "bc");
1208 compilerOutputRemove("myOutputFileName", "spv");
1209 }
1210
TEST(OfflineCompilerTest,GivenKernelWhenNoCharAfterKernelSourceThenBuildWithSuccess)1211 TEST(OfflineCompilerTest, GivenKernelWhenNoCharAfterKernelSourceThenBuildWithSuccess) {
1212 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1213 ASSERT_NE(nullptr, mockOfflineCompiler);
1214
1215 auto retVal = mockOfflineCompiler->buildSourceCode();
1216 EXPECT_EQ(CL_INVALID_PROGRAM, retVal);
1217
1218 std::vector<std::string> argv = {
1219 "ocloc",
1220 "-file",
1221 "test_files/emptykernel.cl",
1222 "-device",
1223 gEnvironment->devicePrefix.c_str()};
1224
1225 retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1226 EXPECT_EQ(CL_SUCCESS, retVal);
1227
1228 retVal = mockOfflineCompiler->buildSourceCode();
1229 EXPECT_EQ(CL_SUCCESS, retVal);
1230 }
1231
TEST(OfflineCompilerTest,WhenGeneratingElfBinaryThenBinaryIsCreated)1232 TEST(OfflineCompilerTest, WhenGeneratingElfBinaryThenBinaryIsCreated) {
1233 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1234 ASSERT_NE(nullptr, mockOfflineCompiler);
1235
1236 auto retVal = mockOfflineCompiler->generateElfBinary();
1237 EXPECT_FALSE(retVal);
1238
1239 iOpenCL::SProgramBinaryHeader binHeader;
1240 memset(&binHeader, 0, sizeof(binHeader));
1241 binHeader.Magic = iOpenCL::MAGIC_CL;
1242 binHeader.Version = iOpenCL::CURRENT_ICBE_VERSION - 3;
1243 binHeader.Device = DEFAULT_PLATFORM::hwInfo.platform.eRenderCoreFamily;
1244 binHeader.GPUPointerSizeInBytes = 8;
1245 binHeader.NumberOfKernels = 0;
1246 binHeader.SteppingId = 0;
1247 binHeader.PatchListSize = 0;
1248 size_t binSize = sizeof(iOpenCL::SProgramBinaryHeader);
1249
1250 mockOfflineCompiler->storeGenBinary(&binHeader, binSize);
1251
1252 EXPECT_TRUE(mockOfflineCompiler->elfBinary.empty());
1253
1254 retVal = mockOfflineCompiler->generateElfBinary();
1255 EXPECT_TRUE(retVal);
1256
1257 EXPECT_FALSE(mockOfflineCompiler->elfBinary.empty());
1258 }
1259
TEST(OfflineCompilerTest,givenLlvmInputOptionPassedWhenCmdLineParsedThenInputFileLlvmIsSetTrue)1260 TEST(OfflineCompilerTest, givenLlvmInputOptionPassedWhenCmdLineParsedThenInputFileLlvmIsSetTrue) {
1261 std::vector<std::string> argv = {
1262 "ocloc",
1263 "-llvm_input"};
1264
1265 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1266 ASSERT_NE(nullptr, mockOfflineCompiler);
1267
1268 testing::internal::CaptureStdout();
1269 mockOfflineCompiler->parseCommandLine(argv.size(), argv);
1270 std::string output = testing::internal::GetCapturedStdout();
1271
1272 EXPECT_NE(0u, output.size());
1273
1274 bool llvmFileOption = mockOfflineCompiler->inputFileLlvm;
1275 EXPECT_TRUE(llvmFileOption);
1276 }
1277
TEST(OfflineCompilerTest,givenDefaultOfflineCompilerObjectWhenNoOptionsAreChangedThenLlvmInputFileIsFalse)1278 TEST(OfflineCompilerTest, givenDefaultOfflineCompilerObjectWhenNoOptionsAreChangedThenLlvmInputFileIsFalse) {
1279 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1280 ASSERT_NE(nullptr, mockOfflineCompiler);
1281
1282 bool llvmFileOption = mockOfflineCompiler->inputFileLlvm;
1283 EXPECT_FALSE(llvmFileOption);
1284 }
1285
TEST(OfflineCompilerTest,givenSpirvInputOptionPassedWhenCmdLineParsedThenInputFileSpirvIsSetTrue)1286 TEST(OfflineCompilerTest, givenSpirvInputOptionPassedWhenCmdLineParsedThenInputFileSpirvIsSetTrue) {
1287 std::vector<std::string> argv = {"ocloc", "-spirv_input"};
1288
1289 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1290
1291 testing::internal::CaptureStdout();
1292 mockOfflineCompiler->parseCommandLine(argv.size(), argv);
1293 std::string output = testing::internal::GetCapturedStdout();
1294 EXPECT_NE(0u, output.size());
1295
1296 EXPECT_TRUE(mockOfflineCompiler->inputFileSpirV);
1297 }
1298
TEST(OfflineCompilerTest,givenDefaultOfflineCompilerObjectWhenNoOptionsAreChangedThenSpirvInputFileIsFalse)1299 TEST(OfflineCompilerTest, givenDefaultOfflineCompilerObjectWhenNoOptionsAreChangedThenSpirvInputFileIsFalse) {
1300 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1301 EXPECT_FALSE(mockOfflineCompiler->inputFileSpirV);
1302 }
1303
TEST(OfflineCompilerTest,givenIntermediateRepresentationInputWhenBuildSourceCodeIsCalledThenProperTranslationContextIsUsed)1304 TEST(OfflineCompilerTest, givenIntermediateRepresentationInputWhenBuildSourceCodeIsCalledThenProperTranslationContextIsUsed) {
1305 MockOfflineCompiler mockOfflineCompiler;
1306 std::vector<std::string> argv = {
1307 "ocloc",
1308 "-file",
1309 "test_files/emptykernel.cl",
1310 "-device",
1311 gEnvironment->devicePrefix.c_str()};
1312
1313 testing::internal::CaptureStdout();
1314 auto retVal = mockOfflineCompiler.initialize(argv.size(), argv);
1315 auto mockIgcOclDeviceCtx = new NEO::MockIgcOclDeviceCtx();
1316 mockOfflineCompiler.igcDeviceCtx = CIF::RAII::Pack<IGC::IgcOclDeviceCtxTagOCL>(mockIgcOclDeviceCtx);
1317 ASSERT_EQ(CL_SUCCESS, retVal);
1318
1319 mockOfflineCompiler.inputFileSpirV = true;
1320 retVal = mockOfflineCompiler.buildSourceCode();
1321 EXPECT_EQ(CL_SUCCESS, retVal);
1322 ASSERT_EQ(1U, mockIgcOclDeviceCtx->requestedTranslationCtxs.size());
1323 NEO::MockIgcOclDeviceCtx::TranslationOpT expectedTranslation = {IGC::CodeType::spirV, IGC::CodeType::oclGenBin};
1324 ASSERT_EQ(expectedTranslation, mockIgcOclDeviceCtx->requestedTranslationCtxs[0]);
1325
1326 mockOfflineCompiler.inputFileSpirV = false;
1327 mockOfflineCompiler.inputFileLlvm = true;
1328 mockIgcOclDeviceCtx->requestedTranslationCtxs.clear();
1329 retVal = mockOfflineCompiler.buildSourceCode();
1330
1331 ASSERT_EQ(mockOfflineCompiler.irBinarySize, mockOfflineCompiler.sourceCode.size());
1332 EXPECT_EQ(0, memcmp(mockOfflineCompiler.irBinary, mockOfflineCompiler.sourceCode.data(), mockOfflineCompiler.sourceCode.size()));
1333 EXPECT_FALSE(mockOfflineCompiler.isSpirV);
1334 EXPECT_EQ(CL_SUCCESS, retVal);
1335 ASSERT_EQ(1U, mockIgcOclDeviceCtx->requestedTranslationCtxs.size());
1336 expectedTranslation = {IGC::CodeType::llvmBc, IGC::CodeType::oclGenBin};
1337 ASSERT_EQ(expectedTranslation, mockIgcOclDeviceCtx->requestedTranslationCtxs[0]);
1338
1339 testing::internal::GetCapturedStdout();
1340 }
1341
TEST(OfflineCompilerTest,givenBinaryInputThenDontTruncateSourceAtFirstZero)1342 TEST(OfflineCompilerTest, givenBinaryInputThenDontTruncateSourceAtFirstZero) {
1343 std::vector<std::string> argvLlvm = {"ocloc", "-llvm_input", "-file", "test_files/binary_with_zeroes",
1344 "-device", gEnvironment->devicePrefix.c_str()};
1345 auto mockOfflineCompiler = std::make_unique<MockOfflineCompiler>();
1346 mockOfflineCompiler->initialize(argvLlvm.size(), argvLlvm);
1347 EXPECT_LT(0U, mockOfflineCompiler->sourceCode.size());
1348
1349 std::vector<std::string> argvSpirV = {"ocloc", "-spirv_input", "-file", "test_files/binary_with_zeroes",
1350 "-device", gEnvironment->devicePrefix.c_str()};
1351 mockOfflineCompiler = std::make_unique<MockOfflineCompiler>();
1352 mockOfflineCompiler->initialize(argvSpirV.size(), argvSpirV);
1353 EXPECT_LT(0U, mockOfflineCompiler->sourceCode.size());
1354 }
1355
TEST(OfflineCompilerTest,givenSpirvInputFileWhenCmdLineHasOptionsThenCorrectOptionsArePassedToCompiler)1356 TEST(OfflineCompilerTest, givenSpirvInputFileWhenCmdLineHasOptionsThenCorrectOptionsArePassedToCompiler) {
1357 char data[] = {1, 2, 3, 4, 5, 6, 7, 8};
1358 MockCompilerDebugVars igcDebugVars(gEnvironment->igcDebugVars);
1359 igcDebugVars.binaryToReturn = data;
1360 igcDebugVars.binaryToReturnSize = sizeof(data);
1361
1362 NEO::setIgcDebugVars(igcDebugVars);
1363
1364 MockOfflineCompiler mockOfflineCompiler;
1365 std::vector<std::string> argv = {
1366 "ocloc",
1367 "-file",
1368 "test_files/emptykernel.cl",
1369 "-spirv_input",
1370 "-device",
1371 gEnvironment->devicePrefix.c_str(),
1372 "-options",
1373 "test_options_passed"};
1374
1375 auto retVal = mockOfflineCompiler.initialize(argv.size(), argv);
1376 auto mockIgcOclDeviceCtx = new NEO::MockIgcOclDeviceCtx();
1377 mockOfflineCompiler.igcDeviceCtx = CIF::RAII::Pack<IGC::IgcOclDeviceCtxTagOCL>(mockIgcOclDeviceCtx);
1378 ASSERT_EQ(CL_SUCCESS, retVal);
1379
1380 mockOfflineCompiler.inputFileSpirV = true;
1381
1382 retVal = mockOfflineCompiler.buildSourceCode();
1383 EXPECT_EQ(CL_SUCCESS, retVal);
1384
1385 EXPECT_STREQ("test_options_passed", mockOfflineCompiler.options.c_str());
1386 NEO::setIgcDebugVars(gEnvironment->igcDebugVars);
1387 }
1388
TEST(OfflineCompilerTest,givenOutputFileOptionWhenSourceIsCompiledThenOutputFileHasCorrectName)1389 TEST(OfflineCompilerTest, givenOutputFileOptionWhenSourceIsCompiledThenOutputFileHasCorrectName) {
1390 std::vector<std::string> argv = {
1391 "ocloc",
1392 "-file",
1393 "test_files/copybuffer.cl",
1394 "-output",
1395 "myOutputFileName",
1396 "-device",
1397 gEnvironment->devicePrefix.c_str()};
1398
1399 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1400 ASSERT_NE(nullptr, mockOfflineCompiler);
1401
1402 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1403 EXPECT_EQ(CL_SUCCESS, retVal);
1404
1405 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1406 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bin"));
1407 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "gen"));
1408
1409 retVal = mockOfflineCompiler->build();
1410 EXPECT_EQ(CL_SUCCESS, retVal);
1411
1412 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1413 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "bin"));
1414 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "gen"));
1415
1416 compilerOutputRemove("myOutputFileName", "bc");
1417 compilerOutputRemove("myOutputFileName", "spv");
1418 compilerOutputRemove("myOutputFileName", "bin");
1419 compilerOutputRemove("myOutputFileName", "gen");
1420 }
1421
TEST(OfflineCompilerTest,givenDebugDataAvailableWhenSourceIsBuiltThenDebugDataFileIsCreated)1422 TEST(OfflineCompilerTest, givenDebugDataAvailableWhenSourceIsBuiltThenDebugDataFileIsCreated) {
1423 std::vector<std::string> argv = {
1424 "ocloc",
1425 "-file",
1426 "test_files/copybuffer.cl",
1427 "-output",
1428 "myOutputFileName",
1429 "-device",
1430 gEnvironment->devicePrefix.c_str()};
1431
1432 char debugData[10];
1433 MockCompilerDebugVars igcDebugVars(gEnvironment->igcDebugVars);
1434 igcDebugVars.debugDataToReturn = debugData;
1435 igcDebugVars.debugDataToReturnSize = sizeof(debugData);
1436
1437 NEO::setIgcDebugVars(igcDebugVars);
1438
1439 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1440 ASSERT_NE(nullptr, mockOfflineCompiler);
1441
1442 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1443 EXPECT_EQ(CL_SUCCESS, retVal);
1444
1445 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1446 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "bin"));
1447 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "gen"));
1448 EXPECT_FALSE(compilerOutputExists("myOutputFileName", "dbg"));
1449
1450 retVal = mockOfflineCompiler->build();
1451 EXPECT_EQ(CL_SUCCESS, retVal);
1452
1453 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "bc") || compilerOutputExists("myOutputFileName", "spv"));
1454 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "bin"));
1455 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "gen"));
1456 EXPECT_TRUE(compilerOutputExists("myOutputFileName", "dbg"));
1457
1458 compilerOutputRemove("myOutputFileName", "bc");
1459 compilerOutputRemove("myOutputFileName", "spv");
1460 compilerOutputRemove("myOutputFileName", "bin");
1461 compilerOutputRemove("myOutputFileName", "gen");
1462 compilerOutputRemove("myOutputFileName", "dbg");
1463
1464 NEO::setIgcDebugVars(gEnvironment->igcDebugVars);
1465 }
1466
TEST(OfflineCompilerTest,givenInternalOptionsWhenCmdLineParsedThenOptionsAreAppendedToInternalOptionsString)1467 TEST(OfflineCompilerTest, givenInternalOptionsWhenCmdLineParsedThenOptionsAreAppendedToInternalOptionsString) {
1468 std::vector<std::string> argv = {
1469 "ocloc",
1470 "-internal_options",
1471 "myInternalOptions"};
1472
1473 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1474 ASSERT_NE(nullptr, mockOfflineCompiler);
1475
1476 testing::internal::CaptureStdout();
1477 mockOfflineCompiler->parseCommandLine(argv.size(), argv);
1478 std::string output = testing::internal::GetCapturedStdout();
1479
1480 EXPECT_NE(0u, output.size());
1481
1482 std::string internalOptions = mockOfflineCompiler->internalOptions;
1483 EXPECT_THAT(internalOptions, ::testing::HasSubstr(std::string("myInternalOptions")));
1484 }
1485
TEST(OfflineCompilerTest,givenInputOptionsAndInternalOptionsFilesWhenOfflineCompilerIsInitializedThenCorrectOptionsAreSetAndRemainAfterBuild)1486 TEST(OfflineCompilerTest, givenInputOptionsAndInternalOptionsFilesWhenOfflineCompilerIsInitializedThenCorrectOptionsAreSetAndRemainAfterBuild) {
1487 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1488 ASSERT_NE(nullptr, mockOfflineCompiler);
1489
1490 ASSERT_TRUE(fileExists("test_files/shouldfail_options.txt"));
1491 ASSERT_TRUE(fileExists("test_files/shouldfail_internal_options.txt"));
1492
1493 std::vector<std::string> argv = {
1494 "ocloc",
1495 "-q",
1496 "-file",
1497 "test_files/shouldfail.cl",
1498 "-device",
1499 gEnvironment->devicePrefix.c_str()};
1500
1501 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1502 EXPECT_EQ(CL_SUCCESS, retVal);
1503
1504 auto &options = mockOfflineCompiler->options;
1505 auto &internalOptions = mockOfflineCompiler->internalOptions;
1506 EXPECT_STREQ(options.c_str(), "-shouldfailOptions");
1507 EXPECT_TRUE(internalOptions.find("-shouldfailInternalOptions") != std::string::npos);
1508
1509 mockOfflineCompiler->build();
1510
1511 EXPECT_STREQ(options.c_str(), "-shouldfailOptions");
1512 EXPECT_TRUE(internalOptions.find("-shouldfailInternalOptions") != std::string::npos);
1513 }
1514
TEST(OfflineCompilerTest,givenInputOptionsFileWithSpecialCharsWhenOfflineCompilerIsInitializedThenCorrectOptionsAreSet)1515 TEST(OfflineCompilerTest, givenInputOptionsFileWithSpecialCharsWhenOfflineCompilerIsInitializedThenCorrectOptionsAreSet) {
1516 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1517 ASSERT_NE(nullptr, mockOfflineCompiler);
1518
1519 ASSERT_TRUE(fileExists("test_files/simple_kernels_opts_options.txt"));
1520
1521 std::vector<std::string> argv = {
1522 "ocloc",
1523 "-q",
1524 "-file",
1525 "test_files/simple_kernels_opts.cl",
1526 "-device",
1527 gEnvironment->devicePrefix.c_str()};
1528
1529 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1530 EXPECT_EQ(CL_SUCCESS, retVal);
1531
1532 auto &options = mockOfflineCompiler->options;
1533 EXPECT_STREQ(options.c_str(), "-cl-opt-disable -DDEF_WAS_SPECIFIED=1 -DARGS=\", const __global int *arg1, float arg2, const __global int *arg3, float arg4\"");
1534 }
1535
TEST(OfflineCompilerTest,givenInputOptionsAndOclockOptionsFileWithForceStosOptWhenOfflineCompilerIsInitializedThenCompilerOptionGreaterThan4gbBuffersRequiredIsNotApplied)1536 TEST(OfflineCompilerTest, givenInputOptionsAndOclockOptionsFileWithForceStosOptWhenOfflineCompilerIsInitializedThenCompilerOptionGreaterThan4gbBuffersRequiredIsNotApplied) {
1537 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1538 ASSERT_NE(nullptr, mockOfflineCompiler);
1539
1540 ASSERT_TRUE(fileExists("test_files/stateful_copy_buffer_ocloc_options.txt"));
1541
1542 std::vector<std::string> argv = {
1543 "ocloc",
1544 "-q",
1545 "-file",
1546 "test_files/stateful_copy_buffer.cl",
1547 "-device",
1548 gEnvironment->devicePrefix.c_str()};
1549
1550 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1551 EXPECT_EQ(CL_SUCCESS, retVal);
1552
1553 mockOfflineCompiler->build();
1554
1555 auto &internalOptions = mockOfflineCompiler->internalOptions;
1556 size_t found = internalOptions.find(NEO::CompilerOptions::greaterThan4gbBuffersRequired.data());
1557 EXPECT_EQ(std::string::npos, found);
1558 }
1559
TEST(OfflineCompilerTest,givenNonExistingFilenameWhenUsedToReadOptionsThenReadOptionsFromFileReturnsFalse)1560 TEST(OfflineCompilerTest, givenNonExistingFilenameWhenUsedToReadOptionsThenReadOptionsFromFileReturnsFalse) {
1561 std::string options;
1562 std::string file("non_existing_file");
1563 ASSERT_FALSE(fileExists(file.c_str()));
1564
1565 auto helper = std::make_unique<OclocArgHelper>();
1566 bool result = OfflineCompiler::readOptionsFromFile(options, file, helper.get());
1567
1568 EXPECT_FALSE(result);
1569 }
1570
1571 class MyOclocArgHelper : public OclocArgHelper {
1572 public:
loadDataFromFile(const std::string & filename,size_t & retSize)1573 std::unique_ptr<char[]> loadDataFromFile(const std::string &filename, size_t &retSize) override {
1574 auto file = std::make_unique<char[]>(fileContent.size() + 1);
1575 strcpy_s(file.get(), fileContent.size() + 1, fileContent.c_str());
1576 retSize = fileContent.size();
1577 return file;
1578 }
1579
fileExists(const std::string & filename) const1580 bool fileExists(const std::string &filename) const override {
1581 return true;
1582 }
1583
1584 std::string fileContent;
1585 };
1586
TEST(OfflineCompilerTest,givenEmptyFileWhenReadOptionsFromFileThenSuccessIsReturned)1587 TEST(OfflineCompilerTest, givenEmptyFileWhenReadOptionsFromFileThenSuccessIsReturned) {
1588 std::string options;
1589 std::string filename("non_existing_file");
1590
1591 auto helper = std::make_unique<MyOclocArgHelper>();
1592 helper->fileContent = "";
1593
1594 EXPECT_TRUE(OfflineCompiler::readOptionsFromFile(options, filename, helper.get()));
1595 EXPECT_TRUE(options.empty());
1596 }
1597
TEST(OfflineCompilerTest,givenNoCopyrightsWhenReadOptionsFromFileThenSuccessIsReturned)1598 TEST(OfflineCompilerTest, givenNoCopyrightsWhenReadOptionsFromFileThenSuccessIsReturned) {
1599 std::string options;
1600 std::string filename("non_existing_file");
1601
1602 auto helper = std::make_unique<MyOclocArgHelper>();
1603 helper->fileContent = "-dummy_option";
1604 EXPECT_TRUE(OfflineCompiler::readOptionsFromFile(options, filename, helper.get()));
1605 EXPECT_STREQ(helper->fileContent.c_str(), options.c_str());
1606 }
1607
TEST(OfflineCompilerTest,givenEmptyDirectoryWhenGenerateFilePathIsCalledThenTrailingSlashIsNotAppended)1608 TEST(OfflineCompilerTest, givenEmptyDirectoryWhenGenerateFilePathIsCalledThenTrailingSlashIsNotAppended) {
1609 std::string path = generateFilePath("", "a", "b");
1610 EXPECT_STREQ("ab", path.c_str());
1611 }
1612
TEST(OfflineCompilerTest,givenNonEmptyDirectoryWithTrailingSlashWhenGenerateFilePathIsCalledThenAdditionalTrailingSlashIsNotAppended)1613 TEST(OfflineCompilerTest, givenNonEmptyDirectoryWithTrailingSlashWhenGenerateFilePathIsCalledThenAdditionalTrailingSlashIsNotAppended) {
1614 std::string path = generateFilePath("d/", "a", "b");
1615 EXPECT_STREQ("d/ab", path.c_str());
1616 }
1617
TEST(OfflineCompilerTest,givenNonEmptyDirectoryWithoutTrailingSlashWhenGenerateFilePathIsCalledThenTrailingSlashIsAppended)1618 TEST(OfflineCompilerTest, givenNonEmptyDirectoryWithoutTrailingSlashWhenGenerateFilePathIsCalledThenTrailingSlashIsAppended) {
1619 std::string path = generateFilePath("d", "a", "b");
1620 EXPECT_STREQ("d/ab", path.c_str());
1621 }
1622
TEST(OfflineCompilerTest,givenSpirvPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned)1623 TEST(OfflineCompilerTest, givenSpirvPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned) {
1624 MockOfflineCompiler compiler;
1625 compiler.isSpirV = true;
1626 compiler.outputDirectory = "d";
1627 std::string path = compiler.generateFilePathForIr("a");
1628 EXPECT_STREQ("d/a.spv", path.c_str());
1629 }
1630
TEST(OfflineCompilerTest,givenLlvmBcPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned)1631 TEST(OfflineCompilerTest, givenLlvmBcPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned) {
1632 MockOfflineCompiler compiler;
1633 compiler.isSpirV = false;
1634 compiler.outputDirectory = "d";
1635 std::string path = compiler.generateFilePathForIr("a");
1636 EXPECT_STREQ("d/a.bc", path.c_str());
1637 }
1638
TEST(OfflineCompilerTest,givenLlvmTextPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned)1639 TEST(OfflineCompilerTest, givenLlvmTextPathWhenGenerateFilePathForIrIsCalledThenProperExtensionIsReturned) {
1640 MockOfflineCompiler compiler;
1641 compiler.isSpirV = false;
1642 compiler.useLlvmText = true;
1643 compiler.outputDirectory = "d";
1644 std::string path = compiler.generateFilePathForIr("a");
1645 EXPECT_STREQ("d/a.ll", path.c_str());
1646
1647 compiler.isSpirV = true;
1648 path = compiler.generateFilePathForIr("a");
1649 EXPECT_STREQ("d/a.ll", path.c_str());
1650 }
1651
TEST(OfflineCompilerTest,givenDisabledOptsSuffixWhenGenerateOptsSuffixIsCalledThenEmptyStringIsReturned)1652 TEST(OfflineCompilerTest, givenDisabledOptsSuffixWhenGenerateOptsSuffixIsCalledThenEmptyStringIsReturned) {
1653 MockOfflineCompiler compiler;
1654 compiler.options = "A B C";
1655 compiler.useOptionsSuffix = false;
1656 std::string suffix = compiler.generateOptsSuffix();
1657 EXPECT_STREQ("", suffix.c_str());
1658 }
1659
TEST(OfflineCompilerTest,givenEnabledOptsSuffixWhenGenerateOptsSuffixIsCalledThenEscapedStringIsReturned)1660 TEST(OfflineCompilerTest, givenEnabledOptsSuffixWhenGenerateOptsSuffixIsCalledThenEscapedStringIsReturned) {
1661 MockOfflineCompiler compiler;
1662 compiler.options = "A B C";
1663 compiler.useOptionsSuffix = true;
1664 std::string suffix = compiler.generateOptsSuffix();
1665 EXPECT_STREQ("A_B_C", suffix.c_str());
1666 }
TEST(OfflineCompilerTest,givenCompilerWhenBuildSourceCodeFailsThenGenerateElfBinaryAndWriteOutAllFilesAreCalled)1667 TEST(OfflineCompilerTest, givenCompilerWhenBuildSourceCodeFailsThenGenerateElfBinaryAndWriteOutAllFilesAreCalled) {
1668 MockOfflineCompiler compiler;
1669 compiler.overrideBuildSourceCodeStatus = true;
1670
1671 auto expectedError = OfflineCompiler::ErrorCode::BUILD_PROGRAM_FAILURE;
1672 compiler.buildSourceCodeStatus = expectedError;
1673
1674 EXPECT_EQ(0u, compiler.generateElfBinaryCalled);
1675 EXPECT_EQ(0u, compiler.writeOutAllFilesCalled);
1676
1677 auto status = compiler.build();
1678 EXPECT_EQ(expectedError, status);
1679
1680 EXPECT_EQ(1u, compiler.generateElfBinaryCalled);
1681 EXPECT_EQ(1u, compiler.writeOutAllFilesCalled);
1682 }
1683
TEST(OfflineCompilerTest,givenDeviceSpecificKernelFileWhenCompilerIsInitializedThenOptionsAreReadFromFile)1684 TEST(OfflineCompilerTest, givenDeviceSpecificKernelFileWhenCompilerIsInitializedThenOptionsAreReadFromFile) {
1685 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1686 ASSERT_NE(nullptr, mockOfflineCompiler);
1687 const char *kernelFileName = "test_files/kernel_for_specific_device.skl";
1688 const char *optionsFileName = "test_files/kernel_for_specific_device_options.txt";
1689
1690 ASSERT_TRUE(fileExists(kernelFileName));
1691 ASSERT_TRUE(fileExists(optionsFileName));
1692
1693 std::vector<std::string> argv = {
1694 "ocloc",
1695 "-q",
1696 "-file",
1697 kernelFileName,
1698 "-device",
1699 gEnvironment->devicePrefix.c_str()};
1700
1701 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1702 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1703 EXPECT_STREQ("-cl-opt-disable", mockOfflineCompiler->options.c_str());
1704 }
1705
TEST(OfflineCompilerTest,givenHexadecimalRevisionIdWhenCompilerIsInitializedThenPassItToHwInfo)1706 TEST(OfflineCompilerTest, givenHexadecimalRevisionIdWhenCompilerIsInitializedThenPassItToHwInfo) {
1707 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1708 ASSERT_NE(nullptr, mockOfflineCompiler);
1709
1710 std::vector<std::string> argv = {
1711 "ocloc",
1712 "-q",
1713 "-file",
1714 "test_files/copybuffer.cl",
1715 "-device",
1716 gEnvironment->devicePrefix.c_str(),
1717 "-revision_id",
1718 "0x11"};
1719
1720 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1721 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1722 EXPECT_EQ(mockOfflineCompiler->hwInfo.platform.usRevId, 17);
1723 }
1724
TEST(OfflineCompilerTest,givenDebugVariableSetWhenInitializingThenOverrideRevision)1725 TEST(OfflineCompilerTest, givenDebugVariableSetWhenInitializingThenOverrideRevision) {
1726 DebugManagerStateRestore stateRestore;
1727 DebugManager.flags.OverrideRevision.set(123);
1728
1729 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1730 ASSERT_NE(nullptr, mockOfflineCompiler);
1731
1732 std::vector<std::string> argv = {
1733 "ocloc",
1734 "-q",
1735 "-file",
1736 "test_files/copybuffer.cl",
1737 "-device",
1738 gEnvironment->devicePrefix.c_str(),
1739 "-revision_id",
1740 "0x11"};
1741
1742 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1743 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1744 EXPECT_EQ(mockOfflineCompiler->hwInfo.platform.usRevId, 123);
1745 }
1746
TEST(OfflineCompilerTest,givenDecimalRevisionIdWhenCompilerIsInitializedThenPassItToHwInfo)1747 TEST(OfflineCompilerTest, givenDecimalRevisionIdWhenCompilerIsInitializedThenPassItToHwInfo) {
1748 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1749 ASSERT_NE(nullptr, mockOfflineCompiler);
1750
1751 std::vector<std::string> argv = {
1752 "ocloc",
1753 "-q",
1754 "-file",
1755 "test_files/copybuffer.cl",
1756 "-device",
1757 gEnvironment->devicePrefix.c_str(),
1758 "-revision_id",
1759 "17"};
1760
1761 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1762 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1763 EXPECT_EQ(mockOfflineCompiler->hwInfo.platform.usRevId, 17);
1764 }
1765
TEST(OfflineCompilerTest,givenNoRevisionIdWhenCompilerIsInitializedThenHwInfoHasDefaultRevId)1766 TEST(OfflineCompilerTest, givenNoRevisionIdWhenCompilerIsInitializedThenHwInfoHasDefaultRevId) {
1767 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1768 ASSERT_NE(nullptr, mockOfflineCompiler);
1769
1770 std::vector<std::string> argv = {
1771 "ocloc",
1772 "-q",
1773 "-file",
1774 "test_files/copybuffer.cl",
1775 "-device",
1776 gEnvironment->devicePrefix.c_str()};
1777
1778 mockOfflineCompiler->initHardwareInfo(gEnvironment->devicePrefix.c_str());
1779 auto revId = mockOfflineCompiler->hwInfo.platform.usRevId;
1780 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1781 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1782 EXPECT_EQ(mockOfflineCompiler->hwInfo.platform.usRevId, revId);
1783 }
1784
TEST(OfflineCompilerTest,whenDeviceIsSpecifiedThenDefaultConfigFromTheDeviceIsUsed)1785 TEST(OfflineCompilerTest, whenDeviceIsSpecifiedThenDefaultConfigFromTheDeviceIsUsed) {
1786 auto mockOfflineCompiler = std::unique_ptr<MockOfflineCompiler>(new MockOfflineCompiler());
1787 ASSERT_NE(nullptr, mockOfflineCompiler);
1788
1789 std::vector<std::string> argv = {
1790 "ocloc",
1791 "-q",
1792 "-file",
1793 "test_files/copybuffer.cl",
1794 "-device",
1795 gEnvironment->devicePrefix.c_str()};
1796
1797 int retVal = mockOfflineCompiler->initialize(argv.size(), argv);
1798 EXPECT_EQ(OfflineCompiler::ErrorCode::SUCCESS, retVal);
1799
1800 HardwareInfo hwInfo = mockOfflineCompiler->hwInfo;
1801
1802 uint32_t sliceCount = 2;
1803 uint32_t subSlicePerSliceCount = 4;
1804 uint32_t euPerSubSliceCount = 5;
1805
1806 uint64_t hwInfoConfig = euPerSubSliceCount;
1807 hwInfoConfig |= (static_cast<uint64_t>(subSlicePerSliceCount) << 16);
1808 hwInfoConfig |= (static_cast<uint64_t>(sliceCount) << 32);
1809
1810 setHwInfoValuesFromConfig(hwInfoConfig, hwInfo);
1811
1812 EXPECT_EQ(sliceCount, hwInfo.gtSystemInfo.SliceCount);
1813 EXPECT_EQ(subSlicePerSliceCount * sliceCount, hwInfo.gtSystemInfo.SubSliceCount);
1814 EXPECT_EQ(subSlicePerSliceCount * sliceCount, hwInfo.gtSystemInfo.SubSliceCount);
1815 EXPECT_EQ(euPerSubSliceCount * subSlicePerSliceCount * sliceCount, hwInfo.gtSystemInfo.EUCount);
1816 }
1817
TEST(OclocCompile,whenDetectedPotentialInputTypeMismatchThenEmitsWarning)1818 TEST(OclocCompile, whenDetectedPotentialInputTypeMismatchThenEmitsWarning) {
1819 std::string sourceOclC = "__kernel void k() { }";
1820 std::string sourceLlvmBc = NEO::llvmBcMagic.str();
1821 std::string sourceSpirv = NEO::spirvMagic.str();
1822 std::string sourceSpirvInv = NEO::spirvMagicInv.str();
1823
1824 std::string notSpirvWarning = "Warning : file does not look like spirv bitcode (wrong magic numbers)";
1825 std::string notLlvmBcWarning = "Warning : file does not look like llvm bitcode (wrong magic numbers)";
1826 std::string isSpirvWarning = "Warning : file looks like spirv bitcode (based on magic numbers) - please make sure proper CLI flags are present";
1827 std::string isLlvmBcWarning = "Warning : file looks like llvm bitcode (based on magic numbers) - please make sure proper CLI flags are present";
1828 std::string allWarnings[] = {notSpirvWarning, notLlvmBcWarning, isLlvmBcWarning, isSpirvWarning};
1829
1830 struct Case {
1831 std::string input;
1832 bool isSpirv;
1833 bool isLlvm;
1834 std::string expectedWarning;
1835 };
1836
1837 Case cases[] = {
1838 {sourceOclC, false, false, ""},
1839 {sourceOclC, true, false, notSpirvWarning},
1840 {sourceOclC, false, true, notLlvmBcWarning},
1841
1842 {sourceLlvmBc, false, false, isLlvmBcWarning},
1843 {sourceLlvmBc, true, false, notSpirvWarning},
1844 {sourceLlvmBc, false, true, ""},
1845
1846 {sourceSpirv, false, false, isSpirvWarning},
1847 {sourceSpirv, true, false, ""},
1848 {sourceSpirv, false, true, notLlvmBcWarning},
1849
1850 {sourceSpirvInv, false, false, isSpirvWarning},
1851 {sourceSpirvInv, true, false, ""},
1852 {sourceSpirvInv, false, true, notLlvmBcWarning},
1853 };
1854 {
1855 MockOfflineCompiler ocloc;
1856
1857 std::vector<std::string> argv = {
1858 "ocloc",
1859 "-q",
1860 "-file",
1861 "test_files/copybuffer.cl",
1862 "-device",
1863 gEnvironment->devicePrefix.c_str()};
1864
1865 ocloc.initHardwareInfo(gEnvironment->devicePrefix.c_str());
1866 int retVal = ocloc.initialize(argv.size(), argv);
1867 ASSERT_EQ(0, retVal);
1868
1869 int caseNum = 0;
1870 for (auto &c : cases) {
1871
1872 testing::internal::CaptureStdout();
1873
1874 ocloc.sourceCode = c.input;
1875 ocloc.inputFileLlvm = c.isLlvm;
1876 ocloc.inputFileSpirV = c.isSpirv;
1877 ocloc.build();
1878 auto log = ocloc.argHelper->getPrinterRef().getLog().str();
1879 ocloc.clearLog();
1880
1881 std::string output = testing::internal::GetCapturedStdout();
1882
1883 if (c.expectedWarning.empty()) {
1884 for (auto &w : allWarnings) {
1885 EXPECT_THAT(log.c_str(), testing::Not(testing::HasSubstr(w.c_str()))) << " Case : " << caseNum;
1886 }
1887 } else {
1888 EXPECT_THAT(log.c_str(), testing::HasSubstr(c.expectedWarning.c_str())) << " Case : " << caseNum;
1889 EXPECT_STREQ(log.c_str(), output.c_str());
1890 }
1891 caseNum++;
1892 }
1893 }
1894 }
1895
TEST(OclocCompile,givenCommandLineWithoutDeviceWhenCompilingToSpirvThenSucceedsButUsesEmptyExtensionString)1896 TEST(OclocCompile, givenCommandLineWithoutDeviceWhenCompilingToSpirvThenSucceedsButUsesEmptyExtensionString) {
1897 MockOfflineCompiler ocloc;
1898
1899 std::vector<std::string> argv = {
1900 "ocloc",
1901 "-q",
1902 "-file",
1903 "test_files/copybuffer.cl",
1904 "-spv_only"};
1905
1906 int retVal = ocloc.initialize(argv.size(), argv);
1907 ASSERT_EQ(0, retVal);
1908 retVal = ocloc.build();
1909 EXPECT_EQ(0, retVal);
1910 EXPECT_THAT(ocloc.internalOptions.c_str(), testing::HasSubstr("-ocl-version=300 -cl-ext=-all,+cl_khr_3d_image_writes -D__IMAGE_SUPPORT__=1"));
1911 }
1912
TEST(OclocCompile,givenDeviceAndInternalOptionsOptionWhenCompilingToSpirvThenInternalOptionsAreSetCorrectly)1913 TEST(OclocCompile, givenDeviceAndInternalOptionsOptionWhenCompilingToSpirvThenInternalOptionsAreSetCorrectly) {
1914 MockOfflineCompiler ocloc;
1915
1916 std::vector<std::string> argv = {
1917 "ocloc",
1918 "-q",
1919 "-file",
1920 "test_files/copybuffer.cl",
1921 "-internal_options",
1922 "-cl-ext=+custom_param",
1923 "-device",
1924 gEnvironment->devicePrefix.c_str(),
1925 "-spv_only"};
1926
1927 int retVal = ocloc.initialize(argv.size(), argv);
1928 ASSERT_EQ(0, retVal);
1929 retVal = ocloc.build();
1930 EXPECT_EQ(0, retVal);
1931 std::string regexToMatch = "\\-ocl\\-version=" + std::to_string(DEFAULT_PLATFORM::hwInfo.capabilityTable.clVersionSupport) +
1932 "0 \\-cl\\-ext=\\-all.* \\-cl\\-ext=\\+custom_param";
1933 EXPECT_THAT(ocloc.internalOptions, ::testing::ContainsRegex(regexToMatch));
1934 }
1935
TEST(OclocCompile,givenNoDeviceAndInternalOptionsOptionWhenCompilingToSpirvThenInternalOptionsAreSetCorrectly)1936 TEST(OclocCompile, givenNoDeviceAndInternalOptionsOptionWhenCompilingToSpirvThenInternalOptionsAreSetCorrectly) {
1937 MockOfflineCompiler ocloc;
1938
1939 std::vector<std::string> argv = {
1940 "ocloc",
1941 "-q",
1942 "-file",
1943 "test_files/copybuffer.cl",
1944 "-internal_options",
1945 "-cl-ext=+custom_param",
1946 "-spv_only"};
1947
1948 int retVal = ocloc.initialize(argv.size(), argv);
1949 ASSERT_EQ(0, retVal);
1950 retVal = ocloc.build();
1951 EXPECT_EQ(0, retVal);
1952 EXPECT_THAT(ocloc.internalOptions.c_str(), testing::HasSubstr("-ocl-version=300 -cl-ext=-all,+cl_khr_3d_image_writes -cl-ext=+custom_param"));
1953 }
TEST(OclocCompile,givenPackedDeviceBinaryFormatWhenGeneratingElfBinaryThenItIsReturnedAsItIs)1954 TEST(OclocCompile, givenPackedDeviceBinaryFormatWhenGeneratingElfBinaryThenItIsReturnedAsItIs) {
1955 MockOfflineCompiler ocloc;
1956 ZebinTestData::ValidEmptyProgram zebin;
1957
1958 // genBinary is deleted in ocloc's destructor
1959 ocloc.genBinary = new char[zebin.storage.size()];
1960 ocloc.genBinarySize = zebin.storage.size();
1961 memcpy_s(ocloc.genBinary, ocloc.genBinarySize, zebin.storage.data(), zebin.storage.size());
1962
1963 ASSERT_EQ(true, ocloc.generateElfBinary());
1964 EXPECT_EQ(0, memcmp(zebin.storage.data(), ocloc.elfBinary.data(), zebin.storage.size()));
1965 }
1966
TEST(OclocCompile,givenSpirvInputThenDontGenerateSpirvFile)1967 TEST(OclocCompile, givenSpirvInputThenDontGenerateSpirvFile) {
1968 MockOfflineCompiler ocloc;
1969
1970 std::vector<std::string> argv = {
1971 "ocloc",
1972 "-q",
1973 "-file",
1974 "test_files/binary_with_zeroes",
1975 "-out_dir",
1976 "offline_compiler_test",
1977 "-device",
1978 gEnvironment->devicePrefix.c_str(),
1979 "-spirv_input"};
1980
1981 int retVal = ocloc.initialize(argv.size(), argv);
1982 ASSERT_EQ(0, retVal);
1983 retVal = ocloc.build();
1984 EXPECT_EQ(0, retVal);
1985 EXPECT_TRUE(compilerOutputExists("offline_compiler_test/binary_with_zeroes", "gen"));
1986 EXPECT_TRUE(compilerOutputExists("offline_compiler_test/binary_with_zeroes", "bin"));
1987 EXPECT_FALSE(compilerOutputExists("offline_compiler_test/binary_with_zeroes", "spv"));
1988 }
1989
TEST(OfflineCompilerTest,GivenDebugFlagWhenSetStatelessToStatefullBufferOffsetFlagThenStatelessToStatefullOptimizationIsSetCorrectly)1990 TEST(OfflineCompilerTest, GivenDebugFlagWhenSetStatelessToStatefullBufferOffsetFlagThenStatelessToStatefullOptimizationIsSetCorrectly) {
1991 DebugManagerStateRestore stateRestore;
1992 MockOfflineCompiler mockOfflineCompiler;
1993 {
1994 DebugManager.flags.EnableStatelessToStatefulBufferOffsetOpt.set(0);
1995 mockOfflineCompiler.initHardwareInfo(gEnvironment->devicePrefix.c_str());
1996 mockOfflineCompiler.setStatelessToStatefullBufferOffsetFlag();
1997 std::string internalOptions = mockOfflineCompiler.internalOptions;
1998 size_t found = internalOptions.find(NEO::CompilerOptions::hasBufferOffsetArg.data());
1999 EXPECT_EQ(std::string::npos, found);
2000 }
2001 {
2002 DebugManager.flags.EnableStatelessToStatefulBufferOffsetOpt.set(1);
2003 mockOfflineCompiler.initHardwareInfo(gEnvironment->devicePrefix.c_str());
2004 mockOfflineCompiler.setStatelessToStatefullBufferOffsetFlag();
2005 std::string internalOptions = mockOfflineCompiler.internalOptions;
2006 size_t found = internalOptions.find(NEO::CompilerOptions::hasBufferOffsetArg.data());
2007 EXPECT_NE(std::string::npos, found);
2008 }
2009 }
2010
2011 } // namespace NEO
2012