Skip to content

Commit 84b9701

Browse files
averinevgaverinevg
committed
Remove is_kernel_str
1 parent d087774 commit 84b9701

20 files changed

+48
-115
lines changed

‎src/anyramdb.cpp

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -28,7 +28,6 @@
2828

2929
#include <miopen/errors.hpp>
3030
#include <miopen/logger.hpp>
31-
#include <miopen/md5.hpp>
3231

3332
#include <boost/filesystem/operations.hpp>
3433
#include <boost/filesystem.hpp>

‎src/binary_cache.cpp

Lines changed: 14 additions & 32 deletions
Original file line numberDiff line numberDiff line change
@@ -147,50 +147,36 @@ KDb GetDb(const TargetProperties& target, size_t num_cu)
147147

148148
boost::filesystem::path GetCacheFile(const std::string& device,
149149
const std::string& name,
150-
const std::string& args,
151-
bool is_kernel_str)
150+
const std::string& args)
152151
{
153-
const std::string filename = (is_kernel_str ? miopen::md5(name) : name) + ".o";
152+
const std::string filename = name + ".o";
154153
return GetCachePath(false) / miopen::md5(device + ":" + args) / filename;
155154
}
156155

157156
#if MIOPEN_ENABLE_SQLITE_KERN_CACHE
158-
static inline std::string GetFilenameForInfo2Logging(const bool is_kernel_str,
159-
const std::string& filename,
160-
const std::string& name)
161-
{
162-
if(!miopen::IsLogging(miopen::LoggingLevel::Info2))
163-
return {}; // Used only in MIOPEN_LOG_I2 -- optimize for speed.
164-
if(is_kernel_str)
165-
return filename + " size=" + std::to_string(name.size());
166-
return filename;
167-
}
168-
169157
std::string LoadBinary(const TargetProperties& target,
170158
const size_t num_cu,
171159
const std::string& name,
172-
const std::string& args,
173-
bool is_kernel_str)
160+
const std::string& args)
174161
{
175162
if(miopen::IsCacheDisabled())
176163
return {};
177164

178165
auto db = GetDb(target, num_cu);
179166

180-
const std::string filename = (is_kernel_str ? miopen::md5(name) : name) + ".o";
167+
const std::string filename = name + ".o";
181168
const KernelConfig cfg{filename, args, ""};
182169

183-
const auto verbose_name = GetFilenameForInfo2Logging(is_kernel_str, filename, name);
184-
MIOPEN_LOG_I2("Loading binary for: " << verbose_name << "; args: " << args);
170+
MIOPEN_LOG_I2("Loading binary for: " << filename << "; args: " << args);
185171
auto record = db.FindRecord(cfg);
186172
if(record)
187173
{
188-
MIOPEN_LOG_I2("Successfully loaded binary for: " << verbose_name << "; args: " << args);
174+
MIOPEN_LOG_I2("Successfully loaded binary for: " << filename << "; args: " << args);
189175
return record.get();
190176
}
191177
else
192178
{
193-
MIOPEN_LOG_I2("Unable to load binary for: " << verbose_name << "; args: " << args);
179+
MIOPEN_LOG_I2("Unable to load binary for: " << filename << "; args: " << args);
194180
return {};
195181
}
196182
}
@@ -199,33 +185,30 @@ void SaveBinary(const std::string& hsaco,
199185
const TargetProperties& target,
200186
const std::size_t num_cu,
201187
const std::string& name,
202-
const std::string& args,
203-
bool is_kernel_str)
188+
const std::string& args)
204189
{
205190
if(miopen::IsCacheDisabled())
206191
return;
207192

208193
auto db = GetDb(target, num_cu);
209194

210-
const std::string filename = (is_kernel_str ? miopen::md5(name) : name) + ".o";
195+
const std::string filename = name + ".o";
211196
KernelConfig cfg{filename, args, hsaco};
212197

213-
const auto verbose_name = GetFilenameForInfo2Logging(is_kernel_str, filename, name);
214-
MIOPEN_LOG_I2("Saving binary for: " << verbose_name << "; args: " << args);
198+
MIOPEN_LOG_I2("Saving binary for: " << filename << "; args: " << args);
215199
db.StoreRecord(cfg);
216200
}
217201
#else
218202
boost::filesystem::path LoadBinary(const TargetProperties& target,
219203
const size_t num_cu,
220204
const std::string& name,
221-
const std::string& args,
222-
bool is_kernel_str)
205+
const std::string& args)
223206
{
224207
if(miopen::IsCacheDisabled())
225208
return {};
226209

227210
(void)num_cu;
228-
auto f = GetCacheFile(target.DbId(), name, args, is_kernel_str);
211+
auto f = GetCacheFile(target.DbId(), name, args);
229212
if(boost::filesystem::exists(f))
230213
{
231214
return f.string();
@@ -239,16 +222,15 @@ boost::filesystem::path LoadBinary(const TargetProperties& target,
239222
void SaveBinary(const boost::filesystem::path& binary_path,
240223
const TargetProperties& target,
241224
const std::string& name,
242-
const std::string& args,
243-
bool is_kernel_str)
225+
const std::string& args)
244226
{
245227
if(miopen::IsCacheDisabled())
246228
{
247229
boost::filesystem::remove(binary_path);
248230
}
249231
else
250232
{
251-
auto p = GetCacheFile(target.DbId(), name, args, is_kernel_str);
233+
auto p = GetCacheFile(target.DbId(), name, args);
252234
boost::filesystem::create_directories(p.parent_path());
253235
boost::filesystem::rename(binary_path, p);
254236
}

‎src/db.cpp

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -28,7 +28,6 @@
2828
#include <miopen/errors.hpp>
2929
#include <miopen/lock_file.hpp>
3030
#include <miopen/logger.hpp>
31-
#include <miopen/md5.hpp>
3231

3332
#include <boost/date_time/posix_time/posix_time_types.hpp>
3433
#include <boost/filesystem.hpp>

‎src/hip/handlehip.cpp

Lines changed: 6 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -497,7 +497,6 @@ KernelInvoke Handle::Run(Kernel k) const
497497

498498
Program Handle::LoadProgram(const std::string& program_name,
499499
std::string params,
500-
bool is_kernel_str,
501500
const std::string& kernel_src) const
502501
{
503502
this->impl->set_ctx();
@@ -511,8 +510,7 @@ Program Handle::LoadProgram(const std::string& program_name,
511510
auto hsaco = miopen::LoadBinary(this->GetTargetProperties(),
512511
this->GetMaxComputeUnits(),
513512
program_name,
514-
params,
515-
is_kernel_str);
513+
params);
516514
if(hsaco.empty())
517515
{
518516
const auto arch_target_id = miopen::SplitDelim(arch_name, ':');
@@ -523,8 +521,7 @@ Program Handle::LoadProgram(const std::string& program_name,
523521
hsaco = miopen::LoadBinary(this->GetTargetProperties(),
524522
this->GetMaxComputeUnits(),
525523
program_name,
526-
orig_params + " -mcpu=" + base_arch,
527-
is_kernel_str);
524+
orig_params + " -mcpu=" + base_arch);
528525
}
529526
}
530527

@@ -534,8 +531,8 @@ Program Handle::LoadProgram(const std::string& program_name,
534531
{
535532
CompileTimer ct;
536533
auto p = HIPOCProgram{
537-
program_name, params, is_kernel_str, this->GetTargetProperties(), kernel_src};
538-
ct.Log("Kernel", is_kernel_str ? std::string() : program_name);
534+
program_name, params, this->GetTargetProperties(), kernel_src};
535+
ct.Log("Kernel", program_name);
539536

540537
// Save to cache
541538
#if MIOPEN_ENABLE_SQLITE_KERN_CACHE
@@ -545,15 +542,14 @@ Program Handle::LoadProgram(const std::string& program_name,
545542
this->GetTargetProperties(),
546543
this->GetMaxComputeUnits(),
547544
program_name,
548-
params,
549-
is_kernel_str);
545+
params);
550546
#else
551547
auto path = miopen::GetCachePath(false) / boost::filesystem::unique_path();
552548
if(p.IsCodeObjectInMemory())
553549
miopen::WriteFile(p.GetCodeObjectBlob(), path);
554550
else
555551
boost::filesystem::copy_file(p.GetCodeObjectPathname(), path);
556-
miopen::SaveBinary(path, this->GetTargetProperties(), program_name, params, is_kernel_str);
552+
miopen::SaveBinary(path, this->GetTargetProperties(), program_name, params);
557553
#endif
558554
p.FreeCodeObjectFileStorage();
559555
return p;

‎src/hipoc/hipoc_program.cpp

Lines changed: 3 additions & 9 deletions
Original file line numberDiff line numberDiff line change
@@ -198,12 +198,11 @@ HIPOCProgramImpl::HIPOCProgramImpl(const std::string& program_name, const std::s
198198

199199
HIPOCProgramImpl::HIPOCProgramImpl(const std::string& program_name,
200200
std::string params,
201-
bool is_kernel_str,
202201
const TargetProperties& target_,
203202
const std::string& kernel_src)
204203
: program(program_name), target(target_)
205204
{
206-
BuildCodeObject(params, is_kernel_str, kernel_src);
205+
BuildCodeObject(params, kernel_src);
207206
if(!binary.empty())
208207
{
209208
module = CreateModuleInMem(binary);
@@ -306,18 +305,14 @@ void HIPOCProgramImpl::BuildCodeObjectInMemory(const std::string& params,
306305
#endif // MIOPEN_USE_COMGR
307306

308307
void HIPOCProgramImpl::BuildCodeObject(std::string params,
309-
bool is_kernel_str,
310308
const std::string& kernel_src)
311309
{
312-
std::string filename = is_kernel_str ? "tinygemm.cl" // Fixed name for miopengemm.
313-
: program;
310+
std::string filename = program;
314311
const auto src = [&]() -> std::string {
315312
if(miopen::EndsWith(filename, ".mlir"))
316313
return {}; // MLIR solutions do not use source code.
317314
if(!kernel_src.empty())
318315
return kernel_src;
319-
if(is_kernel_str)
320-
return program;
321316
return GetKernelSrc(program);
322317
}();
323318

@@ -345,11 +340,10 @@ void HIPOCProgramImpl::BuildCodeObject(std::string params,
345340
HIPOCProgram::HIPOCProgram() {}
346341
HIPOCProgram::HIPOCProgram(const std::string& program_name,
347342
std::string params,
348-
bool is_kernel_str,
349343
const TargetProperties& target,
350344
const std::string& kernel_src)
351345
: impl(std::make_shared<HIPOCProgramImpl>(
352-
program_name, params, is_kernel_str, target, kernel_src))
346+
program_name, params, target, kernel_src))
353347
{
354348
}
355349

‎src/include/miopen/binary_cache.hpp

Lines changed: 5 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -38,35 +38,30 @@ bool IsCacheDisabled();
3838

3939
boost::filesystem::path GetCacheFile(const std::string& device,
4040
const std::string& name,
41-
const std::string& args,
42-
bool is_kernel_str);
41+
const std::string& args);
4342

4443
boost::filesystem::path GetCachePath(bool is_system);
4544

4645
#if !MIOPEN_ENABLE_SQLITE_KERN_CACHE
4746
boost::filesystem::path LoadBinary(const TargetProperties& target,
4847
std::size_t num_cu,
4948
const std::string& name,
50-
const std::string& args,
51-
bool is_kernel_str = false);
49+
const std::string& args);
5250
void SaveBinary(const boost::filesystem::path& binary_path,
5351
const TargetProperties& target,
5452
const std::string& name,
55-
const std::string& args,
56-
bool is_kernel_str = false);
53+
const std::string& args);
5754
#else
5855
std::string LoadBinary(const TargetProperties& target,
5956
std::size_t num_cu,
6057
const std::string& name,
61-
const std::string& args,
62-
bool is_kernel_str = false);
58+
const std::string& args);
6359

6460
void SaveBinary(const std::string& hsaco,
6561
const TargetProperties& target,
6662
std::size_t num_cu,
6763
const std::string& name,
68-
const std::string& args,
69-
bool is_kernel_str = false);
64+
const std::string& args);
7065
#endif
7166

7267
} // namespace miopen

‎src/include/miopen/clhelper.hpp

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -45,7 +45,6 @@ ClProgramPtr LoadProgram(cl_context ctx,
4545
const TargetProperties& target,
4646
const std::string& program,
4747
std::string params,
48-
bool is_kernel_str,
4948
const std::string& kernel_src);
5049
void GetProgramBinary(const ClProgramPtr& program, std::string& binary);
5150
void SaveProgramBinary(const ClProgramPtr& program, const std::string& name);

‎src/include/miopen/generic_search.hpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -336,7 +336,7 @@ void CompileAgent(size_t thread_index,
336336
{
337337
if(profile_h.HasProgram(kernel.kernel_file, kernel.comp_options))
338338
continue;
339-
std::ignore = profile_h.LoadProgram(kernel.kernel_file, kernel.comp_options, false, "");
339+
std::ignore = profile_h.LoadProgram(kernel.kernel_file, kernel.comp_options, "");
340340
}
341341
auto tup = std::make_tuple<PerformanceConfig, ConvSolution, bool>(
342342
std::move(current_config), std::move(current_solution), false);

‎src/include/miopen/handle.hpp

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -131,7 +131,6 @@ struct Handle : miopenHandle
131131

132132
Program LoadProgram(const std::string& program_name,
133133
std::string params,
134-
bool is_kernel_str,
135134
const std::string& kernel_src) const;
136135

137136
bool HasProgram(const std::string& program_name, const std::string& params) const;

‎src/include/miopen/hipoc_program.hpp

Lines changed: 0 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -47,7 +47,6 @@ struct HIPOCProgram
4747
/// Other ctors only guarantee to initialize module.
4848
HIPOCProgram(const std::string& program_name,
4949
std::string params,
50-
bool is_kernel_str,
5150
const TargetProperties& target,
5251
const std::string& kernel_src);
5352
HIPOCProgram(const std::string& program_name, const boost::filesystem::path& hsaco);

0 commit comments

Comments
 (0)